Passing a class object to a kernel

Question

Can we pass an object to a kernel function ?

Consider I have a class

class MyClass
{
    public:
        int value;
        float rate;
        MyClass()
        {
            value = 0; rate = 0;
        }
        MyClass(int v,float r)
        {
            value = v; rate = r;
        }
};

and my Kernel takes an array of the objects of MyClass

__global__ void MyKernel(MyClass * mc)
{
    //Some Calculation
}

Can I pass the array ? How to allocate memory ?? Right now I tried with the following code got CudaMemcpy error

cudaError_t cudaStatus;

MyClass darr[10] ;
cudaStatus = cudaMalloc((void**)&darr, size * sizeof(MyClass));

if (cudaStatus != cudaSuccess) {
    fprintf(stderr, "cudaMalloc failed!");
    goto label1;
}

cudaStatus = cudaMemcpy(darr, arr, size * sizeof(MyClass), cudaMemcpyHostToDevice);
//arr is a host array

Answer 1

There are a few problems here, not all directly related to whatever error you are seeing.

Firstly, you will have to define each class method in both the host and device so that the class can be instantiated in both memory spaces (when you do a copy, only the data members of each instance are copied). So your class declaration should look something like this:

class MyClass
{
    public :
        int value;
        float rate;
        __device__ __host__ MyClass()
        {
            value = 0; rate = 0;
        }
        __device__ __host__ MyClass(int v,float r)
        {
            value = v; rate = r;
        }
        __device__ __host__ ~MyClass() {};
}

You then need to correctly allocate the device memory. If you want an array of MyClass with 10 members on the device, allocate and copy it to the device like this:

MyClass arr[10];
MyClass *darr;
const size_t sz = size_t(10) * sizeof(MyClass);
cudaMalloc((void**)&darr, sz);
cudaMemcpy(darr, &arr[0], sz, cudaMemcpyHostToDevice);

[disclaimer: all code written in browser, never complied or tested, use at own risk]

You can then pass darr as an argument to a kernel from the host.