TAPsCUDA_VertexListMSM.cu File Reference

#include "TAPsCUDA_DataPool.cu"
#include "TAPsCUDA_VertexListMSM_Def.cu"

Include dependency graph for TAPsCUDA_VertexListMSM.cu:

This graph shows which files directly or indirectly include this file:

Go to the source code of this file.

Defines
#define	TAPs_CUDA_VERTEX_LIST_MSM_HPP
Functions
BEGIN_NAMESPACE_TAPs__CUDA void	Global__ModelStrand_AdvSim (unsigned int cudaID, unsigned int numOfVertices, unsigned int numOfThreads, float currentTime, float timeStep, int numOfSubSteps, float ptMass, float Ks, float Kd, float Lrest, float *host_vertex_data, float *host_prev_vertex_data)
	For Suture Model (ModelStrand & ModelSuture)
void	Global__ModelStrand_AdvSim_ADVSIM (unsigned int cudaID, unsigned int numOfVertices, unsigned int numOfThreads, float currentTime, float timeStep, int numOfSubSteps, float ptMass, float Ks, float Kd, float Lrest, float *host_vertex_data, float *host_prev_vertex_data, unsigned int *host_sim_flags_data, float *host_pos_constraint_data)
void	Global__SimPropForMultiPartMeshModel_HalfEdge_AdvSim (int numOfVertices, int numOfThreads, float tCurrent, float tNext, float *host_data)
	For Suture Model (ModelStrand & ModelSuture)
void	Global__HETriMeshOneModelMultiParts_AdvSim (unsigned int cudaID, unsigned int numOfVertices, unsigned int numOfThreads, float currentTime, float timeStep, int numOfSubSteps, float ptMass, float Ks, float Kd, float HKs, float HKd, float *host_vertex_data, float *host_prev_vertex_data, float *host_home_vertex_data, int *host_connection_list, unsigned int max_connection_size)
	For SimPropForMultiPartMeshModel_HalfEdge AdvSim Function.
bool	InitailizeDataForVertexList (unsigned int &cudaID, unsigned int numOfVertices, unsigned int max_connection_size, float *vertexList, float *prevVertexList, float *homeVertexList, int *vertexConnectionList)
	For HETriMeshOneModelMultiParts AdvSim Function.
bool	InitailizeDataForSutureModel (unsigned int &cudaID, unsigned int numOfVertices, float *vertexList, float *prevVertexList)
	Initialize CUDA Data for a Suture Model (ModelStrand & ModelSuture)
bool	InitailizeDataForSutureModel_ADVSIM (unsigned int &cudaID, unsigned int numOfVertices, float *vertexList, float *prevVertexList, unsigned int *simFlagsList, float *posConstraintList)

---

Define Documentation

#define TAPs_CUDA_VERTEX_LIST_MSM_HPP

Definition at line 21 of file TAPsCUDA_VertexListMSM.cu.

---

Function Documentation

void Global__HETriMeshOneModelMultiParts_AdvSim	(	unsigned int	cudaID,
		unsigned int	numOfVertices,
		unsigned int	numOfThreads,
		float	currentTime,
		float	timeStep,
		int	numOfSubSteps,
		float	ptMass,
		float	Ks,
		float	Kd,
		float	HKs,
		float	HKd,
		float *	host_vertex_data,
		float *	host_prev_vertex_data,
		float *	host_home_vertex_data,
		int *	host_connection_list,
		unsigned int	max_connection_size
	)

For SimPropForMultiPartMeshModel_HalfEdge AdvSim Function.

For HETriMeshOneModelMultiParts AdvSim Function CUDA Wrapper Function -- For HETriMeshOneModelMultiParts AdvSim Function

Parameters:

cudaID	CUDA ID
numOfVertices	number of vertices
numOfThreads	number of threads
currentTime	current time
timeStep	time step
numOfSubSteps	number of sub-steps
ptMass	point mass
Ks	spring stiffness
Kd	spring damper
HKs	home spring stiffness
HKd	home spring damper
host_vertex_data	host's vertex data
host_prev_vertex_data	host's previous vertex data
host_home_vertex_data	host's home vertex data
host_connection_list	host's connection index list
max_connection_size	maximum connection size

Definition at line 1191 of file TAPsCUDA_VertexListMSM_Def.cu.

Referenced by HETriMeshOneModelMultiParts< T >::AdvanceSimulation().

{
    //printf( "currentTime: %g, timeStep: %g,\n", currentTime, timeStep );
    //printf( "ptMass: %g, Ks: %g, Kd: %g,\n", ptMass, Ks, Kd );

    //printf( "Number of vertices: %d\n", numOfVertices );
    //printf( "Number of threads: %d\n", numOfThreads );
    //printf( "Number of grids: %d\n", numOfVertices/numOfThreads );

    // Allocate device memory for out data
    unsigned int size = sizeof(float) * numOfVertices * 4;
    float * out_data;
    CUDA_SAFE_CALL( cudaMalloc( (void **)&out_data, size ) );

    // Copy Vertex List (and Previous Vertex List) from host data to device data, 
    // since vertices are dynamically changed.
    // While home vertices and connection list are unchanged.
    CUDA_SAFE_CALL( cudaMemcpy( DATA_GlobalPool.DataForVertexList[cudaID]->GetVertexList(), host_vertex_data, size, cudaMemcpyHostToDevice ) );
    // The Previous Vertex List can be saved in CUDA memory.
    // Hence, the Previous Vertex List does not have to be updated.
    //CUDA_SAFE_CALL( cudaMemcpy( DATA_GlobalPool.DataForVertexList[cudaID]->GetPrevVertexList(), host_prev_vertex_data, size, cudaMemcpyHostToDevice ) );

    // Kernel invocation
    unsigned int numOfThreadBlocks = ( numOfVertices + numOfThreads - 1 ) / numOfThreads;
    dim3 Dg( numOfThreadBlocks, 1, 1 );
    dim3 Db( numOfThreads, 1, 1 );
    size_t Ns = 0;
    cudaStream_t S = 0;

    // Bind Textures
    DATA_GlobalPool.DataForVertexList[cudaID]->BindVertexList( numOfVertices );
    DATA_GlobalPool.DataForVertexList[cudaID]->BindPrevVertexList( numOfVertices );
    DATA_GlobalPool.DataForVertexList[cudaID]->BindHomeVertexList( numOfVertices );
    DATA_GlobalPool.DataForVertexList[cudaID]->BindVertexConnectionList( numOfVertices, max_connection_size );

    int CHOICE = 2;

    // HOME VERTEX
    if ( CHOICE == 1 ) {
        // Call the CUDA kernel (Choice 1 -- Home Vertex Position Only)
        Global__HETriMeshOneModelMultiParts_AdvSim_CU<1><<< Dg, Db, Ns, S >>>( 
            numOfVertices, numOfThreads, 
            currentTime, timeStep, 
            ptMass, Ks, Kd, HKs, HKd, 
            max_connection_size, 
            out_data 
        );
        // Swap pointers for Vertex List and Previous Vertex List
        // So that the Previous Vertex List is remain unchanged in the CUDA memory.
        DATA_GlobalPool.DataForVertexList[cudaID]->SwapVertexList();
    }

    // MASS SPRING SYSTEM
    else if ( CHOICE == 2 ) {
        float subTimeStep = timeStep / numOfSubSteps;
        for ( int i = 0; i < numOfSubSteps; ++i ) {
            //printf( "delta time: %g\n", timeStep/numOfSubSteps );

            // Call the CUDA kernel (Choice 2 -- Mass Spring System)
            Global__HETriMeshOneModelMultiParts_AdvSim_CU<2><<< Dg, Db, Ns, S >>>( 
                numOfVertices, numOfThreads, 
                currentTime, subTimeStep, 
                ptMass, Ks, Kd, HKs, HKd, 
                max_connection_size, 
                out_data 
            );
            // Swap pointers for Vertex List and Previous Vertex List
            // So that the Previous Vertex List is remain unchanged in the CUDA memory.
            DATA_GlobalPool.DataForVertexList[cudaID]->SwapVertexList();

            // Copy data to the device's memory for (current) Vertex List
            CUDA_SAFE_CALL( cudaMemcpy( DATA_GlobalPool.DataForVertexList[cudaID]->GetVertexList(), out_data, size, cudaMemcpyDeviceToDevice ) );
            currentTime += subTimeStep;
        }
    }

    // Unbind Textures
    DATA_GlobalPool.DataForVertexList[cudaID]->UnbindVertexList();
    DATA_GlobalPool.DataForVertexList[cudaID]->UnbindPrevVertexList();
    DATA_GlobalPool.DataForVertexList[cudaID]->UnbindHomeVertexList();
    DATA_GlobalPool.DataForVertexList[cudaID]->UnbindVertexConnectionList();

    // Check if kernel execution generated and error
    CUT_CHECK_ERROR( "Kernel execution failed!" );

    // Copy output data to host data Previous Vertex List
    // The host will have to swap its pointers (for current and previous vertex list)
    if ( CHOICE > 0 ) {
        CUDA_SAFE_CALL( cudaMemcpy( host_prev_vertex_data, out_data, size, cudaMemcpyDeviceToHost ) );
    }

    //DATA_GlobalPool.DataForVertexList[cudaID]->PrintDebug( numOfVertices, host_vertex_data, host_prev_vertex_data, host_home_vertex_data );
    //DATA_GlobalPool.DataForVertexList[cudaID]->PrintDebug_ConnectionList( numOfVertices, max_connection_size, host_connection_list );

    // Free device memory
    CUDA_SAFE_CALL( cudaFree( out_data ) );
}

Here is the caller graph for this function:

BEGIN_NAMESPACE_TAPs__CUDA void Global__ModelStrand_AdvSim	(	unsigned int	cudaID,
		unsigned int	numOfVertices,
		unsigned int	numOfThreads,
		float	currentTime,
		float	timeStep,
		int	numOfSubSteps,
		float	ptMass,
		float	Ks,
		float	Kd,
		float	Lrest,
		float *	host_vertex_data,
		float *	host_prev_vertex_data
	)

For Suture Model (ModelStrand & ModelSuture)

CUDA Wrapper Function -- For ModelStrand AdvanceSimulation Function

Parameters:

cudaID	CUDA ID
numOfVertices	number of vertices
numOfThreads	number of threads per CUDA thread block
currentTime	current time
timeStep	time step
numOfSubSteps	number of sub-steps
ptMass	point mass
Ks	spring stiffness
Kd	spring damper
Lrest	spring rest length
host_vertex_data	host's vertex data
host_prev_vertex_data	host's previous vertex data

Definition at line 630 of file TAPsCUDA_VertexListMSM_Def.cu.

{
    //printf( "FILE:%s LINE:%d Global__ModelStrand_AdvSim\n", __FILE__, __LINE__ );
    //printf( "currentTime: %g, timeStep: %g,\n", currentTime, timeStep );
    //printf( "ptMass: %g, Ks: %g, Kd: %g,\n", ptMass, Ks, Kd );
    //printf( "Number of vertices: %d\n", numOfVertices );
    //printf( "Number of threads: %d\n", numOfThreads );
    //printf( "Number of grids: %d\n", ( numOfVertices + numOfThreads - 1 ) / numOfThreads );

    // Allocate device memory for out data
    unsigned int size = sizeof(float) * numOfVertices * 4;
    float * out_data;
    CUDA_SAFE_CALL( cudaMalloc( (void **)&out_data, size ) );

    // Copy Vertex List (and Previous Vertex List) from host data to device data, 
    // since vertices are dynamically changed.
    // While home vertices and connection list are unchanged.
    CUDA_SAFE_CALL( cudaMemcpy( DATA_GlobalPool.DataForVertexList[cudaID]->GetVertexList(), host_vertex_data, size, cudaMemcpyHostToDevice ) );
    // The Previous Vertex List can be saved in CUDA memory.
    // Hence, the Previous Vertex List does not have to be updated.
    //CUDA_SAFE_CALL( cudaMemcpy( DATA_GlobalPool.DataForVertexList[cudaID]->GetPrevVertexList(), host_prev_vertex_data, size, cudaMemcpyHostToDevice ) );

    // Kernel invocation
    unsigned int numOfThreadBlocks = ( numOfVertices + numOfThreads - 1 ) / numOfThreads;
    dim3 Dg( numOfThreadBlocks, 1, 1 );
    dim3 Db( numOfThreads, 1, 1 );
    size_t Ns = 0;
    cudaStream_t S = 0;

    // Bind Textures
    DATA_GlobalPool.DataForVertexList[cudaID]->BindVertexList( numOfVertices );
    DATA_GlobalPool.DataForVertexList[cudaID]->BindPrevVertexList( numOfVertices );
    //DATA_GlobalPool.DataForVertexList[cudaID]->BindHomeVertexList( numOfVertices );

    //*
    int CHOICE = 1;

    // MASS SPRING SYSTEM
    if ( CHOICE == 1 ) {
        float subTimeStep = timeStep / numOfSubSteps;
        for ( int i = 0; i < numOfSubSteps; ++i ) {
            // Call the CUDA kernel (Choice 1 -- Mass Spring System)
            Global__ModelStrand_AdvSim_CU<1><<< Dg, Db, Ns, S >>>( 
                numOfVertices, numOfThreads, 
                currentTime, subTimeStep, 
                ptMass, Ks, Kd, Lrest, 
                out_data 
            );
            // Swap pointers for Vertex List and Previous Vertex List
            // So that the Previous Vertex List is remain unchanged in the CUDA memory.
            DATA_GlobalPool.DataForVertexList[cudaID]->SwapVertexList();

            // Copy data to the device's memory for (current) Vertex List
            CUDA_SAFE_CALL( cudaMemcpy( DATA_GlobalPool.DataForVertexList[cudaID]->GetVertexList(), out_data, size, cudaMemcpyDeviceToDevice ) );
            currentTime += subTimeStep;
        }
    }
    //*/

    // Unbind Textures
    DATA_GlobalPool.DataForVertexList[cudaID]->UnbindVertexList();
    DATA_GlobalPool.DataForVertexList[cudaID]->UnbindPrevVertexList();
    //DATA_GlobalPool.DataForVertexList[cudaID]->UnbindHomeVertexList();

    // Check if kernel execution generated and error
    CUT_CHECK_ERROR( "Kernel execution failed!" );

    // Copy output data to host data Previous Vertex List
    // The host will have to swap its pointers (for current and previous vertex list)
    if ( CHOICE > 0 ) {
        CUDA_SAFE_CALL( cudaMemcpy( host_prev_vertex_data, out_data, size, cudaMemcpyDeviceToHost ) );
    }

    //DATA_GlobalPool.DataForVertexList[cudaID]->PrintDebug( numOfVertices, host_vertex_data );

    // Free device memory
    CUDA_SAFE_CALL( cudaFree( out_data ) );
}

void Global__ModelStrand_AdvSim_ADVSIM	(	unsigned int	cudaID,
		unsigned int	numOfVertices,
		unsigned int	numOfThreads,
		float	currentTime,
		float	timeStep,
		int	numOfSubSteps,
		float	ptMass,
		float	Ks,
		float	Kd,
		float	Lrest,
		float *	host_vertex_data,
		float *	host_prev_vertex_data,
		unsigned int *	host_sim_flags_data,
		float *	host_pos_constraint_data
	)

CUDA Wrapper Function -- For ModelStrand AdvanceSimulation Function with TAPs_ADVANCED_SIMULATION

Parameters:

cudaID	CUDA ID
numOfVertices	number of vertices
numOfThreads	number of threads per CUDA thread block
currentTime	current time
timeStep	time step
numOfSubSteps	number of sub-steps
ptMass	point mass
Ks	spring stiffness
Kd	spring damper
Lrest	spring rest length
host_vertex_data	host's vertex data
host_prev_vertex_data	host's previous vertex data
host_sim_flags_data	host's simulation flags data
host_pos_constraint_data	host's position constraint data

Definition at line 726 of file TAPsCUDA_VertexListMSM_Def.cu.

{
    printf( "FILE:%s LINE:%d Global__ModelStrand_AdvSim_ADVSIM\n", __FILE__, __LINE__ );
    printf( "currentTime: %g, timeStep: %g,\n", currentTime, timeStep );
    printf( "ptMass: %g, Ks: %g, Kd: %g,\n", ptMass, Ks, Kd );
    printf( "Number of vertices: %d\n", numOfVertices );
    printf( "Number of threads: %d\n", numOfThreads );
    printf( "Number of grids: %d\n", ( numOfVertices + numOfThreads - 1 ) / numOfThreads );

    // Allocate device memory for out data
    unsigned int size = sizeof(float) * numOfVertices * 4;
    float * out_data;
    CUDA_SAFE_CALL( cudaMalloc( (void **)&out_data, size ) );

    // Allocate device memory for out data for changing suture's simulation flags list
    unsigned int size_SimFlagsList = sizeof(unsigned int) * numOfVertices;
    unsigned int * out_data_SimFlagsList;
    CUDA_SAFE_CALL( cudaMalloc( (void **)&out_data_SimFlagsList, size_SimFlagsList ) );

    // Allocate device memory for out data for changing suture's position constraint list
    float * out_data_PosConstraintList;
    CUDA_SAFE_CALL( cudaMalloc( (void **)&out_data_PosConstraintList, size ) );

    // Copy Vertex List (and Previous Vertex List) from host data to device data, 
    // since vertices are dynamically changed.
    // While home vertices and connection list are unchanged.
    CUDA_SAFE_CALL( cudaMemcpy( DATA_GlobalPool.DataForVertexList[cudaID]->GetVertexList(), host_vertex_data, size, cudaMemcpyHostToDevice ) );
    // The Previous Vertex List can be saved in CUDA memory.
    // Hence, the Previous Vertex List does not have to be updated.
    //CUDA_SAFE_CALL( cudaMemcpy( DATA_GlobalPool.DataForVertexList[cudaID]->GetPrevVertexList(), host_prev_vertex_data, size, cudaMemcpyHostToDevice ) );

    CUDA_SAFE_CALL( cudaMemcpy( DATA_GlobalPool.DataForVertexList[cudaID]->GetSimFlagsList(), host_sim_flags_data, sizeof(unsigned int)*numOfVertices, cudaMemcpyHostToDevice ) );
    CUDA_SAFE_CALL( cudaMemcpy( DATA_GlobalPool.DataForVertexList[cudaID]->GetPosConstraintList(), host_pos_constraint_data, size, cudaMemcpyHostToDevice ) );

    // Kernel invocation
    unsigned int numOfThreadBlocks = ( numOfVertices + numOfThreads - 1 ) / numOfThreads;
    dim3 Dg( numOfThreadBlocks, 1, 1 );
    dim3 Db( numOfThreads, 1, 1 );
    size_t Ns = 0;
    cudaStream_t S = 0;

    // Bind Textures
    DATA_GlobalPool.DataForVertexList[cudaID]->BindVertexList( numOfVertices );
    DATA_GlobalPool.DataForVertexList[cudaID]->BindPrevVertexList( numOfVertices );
    DATA_GlobalPool.DataForVertexList[cudaID]->BindSimFlagsList( numOfVertices );
    DATA_GlobalPool.DataForVertexList[cudaID]->BindPosConstraintList( numOfVertices );

    //*
    // Add 10 for TAPs_ADVANCED_SIMULATION
    int CHOICE = 100 + 1;

    // MASS SPRING SYSTEM
    if ( CHOICE == 101 ) {
        float subTimeStep = timeStep / numOfSubSteps;
        for ( int i = 0; i < numOfSubSteps; ++i ) {
            //printf( "delta time: %g\n", timeStep/numOfSubSteps );

            // Call the CUDA kernel (Choice 100+1 -- Mass Spring System)
            // with TAPs_ADVANCED_SIMULATION
            Global__ModelStrand_AdvSim_CU<101><<< Dg, Db, Ns, S >>>( 
                numOfVertices, numOfThreads, 
                currentTime, subTimeStep, 
                ptMass, Ks, Kd, Lrest, 
                out_data 
            );
            // Swap pointers for Vertex List and Previous Vertex List
            // So that the Previous Vertex List is remain unchanged in the CUDA memory.
            DATA_GlobalPool.DataForVertexList[cudaID]->SwapVertexList();

            // Copy data to the device's memory for (current) Vertex List
            CUDA_SAFE_CALL( cudaMemcpy( DATA_GlobalPool.DataForVertexList[cudaID]->GetVertexList(), out_data, size, cudaMemcpyDeviceToDevice ) );
            currentTime += subTimeStep;
        }

        // Call the CUDA kernel to enforce constraints
        // with TAPs_ADVANCED_SIMULATION
        Global__ModelStrand_AdvSim_Enforce_Constraint_CU<101><<< Dg, Db, Ns, S >>>( 
            numOfVertices, numOfThreads, 
            currentTime, subTimeStep, 
            ptMass, Ks, Kd, Lrest, 
            out_data_SimFlagsList, 
            out_data_PosConstraintList 
        );
    }
    //*/

    // Unbind Textures
    DATA_GlobalPool.DataForVertexList[cudaID]->UnbindVertexList();
    DATA_GlobalPool.DataForVertexList[cudaID]->UnbindPrevVertexList();
    DATA_GlobalPool.DataForVertexList[cudaID]->UnbindSimFlagsList();
    DATA_GlobalPool.DataForVertexList[cudaID]->UnbindPosConstraintList();

    // Check if kernel execution generated and error
    CUT_CHECK_ERROR( "Kernel execution failed!" );

    // Copy output data to host data Previous Vertex List
    // The host will have to swap its pointers (for current and previous vertex list)
    if ( CHOICE > 0 ) {
        CUDA_SAFE_CALL( cudaMemcpy( host_prev_vertex_data, out_data, size, cudaMemcpyDeviceToHost ) );
        CUDA_SAFE_CALL( cudaMemcpy( host_sim_flags_data, out_data_SimFlagsList, size_SimFlagsList, cudaMemcpyDeviceToHost ) );
        CUDA_SAFE_CALL( cudaMemcpy( host_pos_constraint_data, out_data_PosConstraintList, size, cudaMemcpyDeviceToHost ) );
    }

    /*
    // DEBUG
    {
        float * data = (float *)malloc( size );
        CUDA_SAFE_CALL( cudaMemcpy( data, out_data, size, cudaMemcpyDeviceToHost ) );
        for ( int i = 0, n = 0; i < 2; ++i, n+=4 ) {
            printf( "Out Vertex# %d: %g %g %g %g \n", i, data[n], data[n+1], data[n+2], data[n+3] );
        }
        for ( int i = numOfVertices-2, n = i*4; i < numOfVertices; ++i, n+=4 ) {
            printf( "Out Vertex# %d: %g %g %g %g \n", i, data[n], data[n+1], data[n+2], data[n+3] );
        }
        free( data );
    }
    */

    DATA_GlobalPool.DataForVertexList[cudaID]->PrintDebug( numOfVertices, host_vertex_data, NULL, NULL, host_sim_flags_data, host_pos_constraint_data );

    // Free device memory
    CUDA_SAFE_CALL( cudaFree( out_data ) );
    CUDA_SAFE_CALL( cudaFree( out_data_SimFlagsList ) );
    CUDA_SAFE_CALL( cudaFree( out_data_PosConstraintList ) );
}

void Global__SimPropForMultiPartMeshModel_HalfEdge_AdvSim	(	int	numOfVertices,
		int	numOfThreads,
		float	tCurrent,
		float	tNext,
		float *	host_data
	)

For Suture Model (ModelStrand & ModelSuture)

For SimPropForMultiPartMeshModel_HalfEdge AdvSim Function CUDA Wrapper Function -- For SimPropForMultiPartMeshModel_HalfEdge AdvSim Function

Parameters:

numOfVertices	number of vertices
numOfThreads	number of threads
tCurrent	current time
tNext	next time
host_data	host data

Definition at line 939 of file TAPsCUDA_VertexListMSM_Def.cu.

Referenced by SimPropForMultiPartMeshModel_HalfEdge< T >::AdvSim().

{
    /*
    // CUDA device properties
    int dev_no;
    CUDA_SAFE_CALL( cudaGetDevice( &dev_no ) );
    cudaDeviceProp * prop;
    CUDA_SAFE_CALL( cudaGetDeviceProperties( prop, dev_no ) );
    //printf( "maxThreadsPerBlock: %d\n", prop->maxThreadsPerBlock );
    */

    // Kernel invocation
    unsigned int numOfThreadBlocks = ( numOfVertices + numOfThreads - 1 ) / numOfThreads;
    dim3 Dg( numOfThreadBlocks, 1, 1 );
    dim3 Db( numOfThreads, 1, 1 );
    size_t Ns = 0;
    cudaStream_t S = 0;

    printf( "Number of vertices: %d\n", numOfVertices );
    printf( "Number of threads: %d\n", numOfThreads );
    printf( "Number of grids: %d\n", numOfVertices/numOfThreads );

    // Allocate device memory
    unsigned int size = sizeof(float) * numOfVertices * 3;
    float * device_data;
    CUDA_SAFE_CALL( cudaMalloc( (void **)&device_data, size ) );
    float * out_data;
    CUDA_SAFE_CALL( cudaMalloc( (void **)&out_data, size ) );

    // Copy host data to device data
    CUDA_SAFE_CALL( cudaMemcpy( device_data, host_data, size, cudaMemcpyHostToDevice ) );

    // Call the CUDA kernel
    Global__SimPropForMultiPartMeshModel_HalfEdge_AdvSim_CU<1><<< Dg, Db, Ns, S >>>
    ( numOfVertices, numOfThreads, currentTime, timeStep, device_data, out_data );

    // Check if kernel execution generated and error
    CUT_CHECK_ERROR( "Kernel execution failed" );

    // Copy output data to host data
    CUDA_SAFE_CALL( cudaMemcpy( host_data, out_data, size, cudaMemcpyDeviceToHost ) );


    // Free device memory
    CUDA_SAFE_CALL( cudaFree( device_data ) );
    CUDA_SAFE_CALL( cudaFree( out_data ) );
}

Here is the caller graph for this function:

bool InitailizeDataForSutureModel	(	unsigned int &	cudaID,
		unsigned int	numOfVertices,
		float *	vertexList,
		float *	prevVertexList
	)

Initialize CUDA Data for a Suture Model (ModelStrand & ModelSuture)

Parameters:

cudaID	CUDA ID for the object
numOfVertices	number of vertices
vertexList	list of xyzw vertices
prevVertexList	list of previous xyzw vertices

Definition at line 60 of file TAPsCUDA_VertexListMSM_Def.cu.

{
    // Here Suture Model uses Vertex List Data
    // However, it has implicit connection (each vertex connect to the previous and next vertices).
    // So the connection list is set to zero.

    // Home vertex is unused for now.  Plan is to use home vertex for sticking suture to a surface.

    // Assign CUDA ID to the object
    // The object has to use cudaID to communicate with the CUDA.
    cudaID = DATA_GlobalPool.SizeOfGlobal_Pool;

    // Allocate CUDA data for the object.
    DATA_Vertex_List * newData = new DATA_Vertex_List( numOfVertices, false, 0, true );
    AddToGlobal_Pool_Of_DATA_Vertex_List( newData );

    // Size of memory for (xyzw) vertices
    int size = numOfVertices * 4 * sizeof(float);

    // Copy Vertex List from host data to device data
    CUDA_SAFE_CALL( cudaMemcpy( newData->GetVertexList(), vertexList, size, cudaMemcpyHostToDevice ) );
    // Copy Previous Vertex List from host data to device data
    CUDA_SAFE_CALL( cudaMemcpy( newData->GetPrevVertexList(), prevVertexList, size, cudaMemcpyHostToDevice ) );
    // Copy Home Vertex List from host data to device data
    //CUDA_SAFE_CALL( cudaMemcpy( newData->GetHomeVertexList(), homeVertexList, size, cudaMemcpyHostToDevice ) );
    // Copy Vertex Connection List from host data to device data

    return true;
}

bool InitailizeDataForSutureModel_ADVSIM	(	unsigned int &	cudaID,
		unsigned int	numOfVertices,
		float *	vertexList,
		float *	prevVertexList,
		unsigned int *	simFlagsList,
		float *	posConstraintList
	)

Initialize CUDA Data for a Suture Model (ModelStrand & ModelSuture) with TAPs_ADVANCED_SIMULATION

Parameters:

cudaID	CUDA ID for the object
numOfVertices	number of vertices
vertexList	list of xyzw vertices
prevVertexList	list of previous xyzw vertices
simFlagsList	list of simulation flags
posConstraintList	list of position constraints

Definition at line 99 of file TAPsCUDA_VertexListMSM_Def.cu.

{
    // Here Suture Model uses Vertex List Data
    // However, it has implicit connection (each vertex connect to the previous and next vertices).
    // So the connection list is set to zero.

    // Home vertex is unused for now.  Plan is to use home vertex for sticking suture to a surface.

    // Assign CUDA ID to the object
    // The object has to use cudaID to communicate with the CUDA.
    cudaID = DATA_GlobalPool.SizeOfGlobal_Pool;

    // Allocate CUDA data for the object.
    DATA_Vertex_List * newData = new DATA_Vertex_List( numOfVertices, false, 0, true );
    AddToGlobal_Pool_Of_DATA_Vertex_List( newData );

    // Size of memory for (xyzw) vertices
    int size = numOfVertices * 4 * sizeof(float);

    // Copy Vertex List from host data to device data
    CUDA_SAFE_CALL( cudaMemcpy( newData->GetVertexList(), vertexList, size, cudaMemcpyHostToDevice ) );
    // Copy Previous Vertex List from host data to device data
    CUDA_SAFE_CALL( cudaMemcpy( newData->GetPrevVertexList(), prevVertexList, size, cudaMemcpyHostToDevice ) );
    // Copy Home Vertex List from host data to device data
    //CUDA_SAFE_CALL( cudaMemcpy( newData->GetHomeVertexList(), homeVertexList, size, cudaMemcpyHostToDevice ) );
    // Copy Vertex Connection List from host data to device data
    //CUDA_SAFE_CALL( cudaMemcpy( newData->GetVertexConnectionList(), vertexConnectionList, numOfVertices*max_connection_size*sizeof(int), cudaMemcpyHostToDevice ) );
    // Copy Simulation Flags List from host data to device data
    CUDA_SAFE_CALL( cudaMemcpy( newData->GetSimFlagsList(), simFlagsList, numOfVertices*sizeof(unsigned int), cudaMemcpyHostToDevice ) );
    // Copy Position Constraint List from host data to device data
    CUDA_SAFE_CALL( cudaMemcpy( newData->GetPosConstraintList(), posConstraintList, size, cudaMemcpyHostToDevice ) );

    return true;
}

bool InitailizeDataForVertexList	(	unsigned int &	cudaID,
		unsigned int	numOfVertices,
		unsigned int	max_connection_size,
		float *	vertexList,
		float *	prevVertexList,
		float *	homeVertexList,
		int *	vertexConnectionList
	)

For HETriMeshOneModelMultiParts AdvSim Function.

Initialize CUDA Data for a Mesh Model by Using Vertex List

Parameters:

cudaID	CUDA ID for the object
numOfVertices	number of vertices
max_connection_size	maximum connection size
vertexList	list of xyz vertices
prevVertexList	list of previous xyzw vertices
homeVertexList	list of xyz home vertices
vertexConnectionList	vertex connection (1D array of 2D array -- vertex id and vertex connections)

Definition at line 21 of file TAPsCUDA_VertexListMSM_Def.cu.

Referenced by HETriMeshOneModelMultiParts< T >::CUDA_Initialize_All().

{
    //printf( "InitailizeDataForVertexList\n" );
    //fflush( stdout );

    // Assign CUDA ID to the object
    // The object has to use cudaID to communicate with the CUDA.
    cudaID = DATA_GlobalPool.SizeOfGlobal_Pool;

    // Allocate CUDA data for the object.
    DATA_Vertex_List * newData = new DATA_Vertex_List( numOfVertices, true, max_connection_size, false );
    AddToGlobal_Pool_Of_DATA_Vertex_List( newData );

    // Size of memory for (xyzw) vertices
    int size = numOfVertices * 4 * sizeof(float);

    // Copy Vertex List from host data to device data
    CUDA_SAFE_CALL( cudaMemcpy( newData->GetVertexList(), vertexList, size, cudaMemcpyHostToDevice ) );
    // Copy Previous Vertex List from host data to device data
    CUDA_SAFE_CALL( cudaMemcpy( newData->GetPrevVertexList(), prevVertexList, size, cudaMemcpyHostToDevice ) );
    // Copy Home Vertex List from host data to device data
    CUDA_SAFE_CALL( cudaMemcpy( newData->GetHomeVertexList(), homeVertexList, size, cudaMemcpyHostToDevice ) );
    // Copy Vertex Connection List from host data to device data
    CUDA_SAFE_CALL( cudaMemcpy( newData->GetVertexConnectionList(), vertexConnectionList, numOfVertices*max_connection_size*sizeof(int), cudaMemcpyHostToDevice ) );

    return true;
}

Here is the caller graph for this function: