Cuda host object to device

Question

I am trying to replicate a big class on my cuda device that contains a lot of variables and methods. I have put the class definition into a .cuh file and I am able to create objects and use them in my device code.

The question now is, is there any way of getting an already existing object from host to device ? I am still using a serial version of my code to read in some geometry and physical data. If it is possible to copy it over to the device without using an intermediate array or so, how does the device handle its size without using sizeof ?

Do I use something like this then for the allocation ?

MyClass *MyObject;
int size = sizeog(MyClass);
cudaMalloc((void**)&MyObject_device, size);
cudaMemCpy(Myobject_device, MyObject, size,   cudaMemcpyHostToDevice);

any advice would very much appreciated.

Answer 1

The CUDA compiler is designed to match the data structure alignment and packing that is used in the host compiler. So you can safely pass an object between device and host and access the members regardless of their alignment requirements.

You can pass objects directly as kernel parameters. For instance:

Host:

MyKernel<<<grid_dim, block_dim>>>(my_object);

Device:

__global__ void MyKernel(MyObject my_object) {

If you need to pass an array of objects, an easy way is to use thrust::device_vector. For instance:

Host:

#include <thrust/device_vector.h>
device_vector<MyObject> my_objects;
...
MyObject* my_objects_d = thrust::raw_pointer_cast(&my_objects[0]);
MyKernel<<<grid_dim, block_dim>>>(my_objects_d);

Device:

__global__ void MyKernel(MyObject* my_objects) {