passing thrust::device_vector to a function by ref

Question

I'm trying to pass device_vector of structures

struct point 
{
    unsigned int x;
    unsigned int y;
}

to a function in a following manner:

void print(thrust::device_vector<point> &points, unsigned int index)
{
    std::cout << points[index].y << points[index].y << std::endl;
}

myvector was initialized properly

print(myvector, 0);

I get following errors:

error: class "thrust::device_reference<point>" has no member "x"
error: class "thrust::device_reference<point>" has no member "y"

What's wrong with it?

Answer 1

From http://thrust.googlecode.com/svn/tags/1.1.0/doc/html/structthrust_1_1device__reference.html :

device_reference acts as a reference to an object stored in device memory. device_reference is not intended to be used directly; rather, this type is the result of deferencing a device_ptr. Similarly, taking the address of a device_reference yields a device_ptr.

Maybe you need something like

(&points[index]).get()->x

instead of

points[index].x

It's a bit ugly, but CUDA needs a mechanism to transport data between RAM and GPU.

Answer 2

Unfortunately, device_reference<T> cannot expose members of T, but it can convert to T.

To implement print, make a temporary copy of each element by converting it to a temporary temp:

void print(thrust::device_vector<point> &points, unsigned int index)
{
    point temp = points[index];
    std::cout << temp.y << temp.y << std::endl;
}

Each time you invoke print, it causes a transfer from GPU to system memory to create the temporary. If you need to print the entire collection of points at once, a more efficient method would copy the entire vector points en masse to a host_vector or std::vector (using thrust::copy) and then iterate through the collection as normal.