一个类的成员函数作为CUDA内核函数

Question

我正在使用CUDA 5.0和一个计算能力为2.1的显卡。

问题非常直接：一个内核函数可以是一个类的一部分吗？ 例如：

class Foo
{
private:
 //...
public:
 __global__ void kernel();
};

__global__ void Foo::kernel()
{
 //implementation here
}

如果不行的话，解决方案就是创建一个作为类成员并在内部调用内核函数的包装函数？

如果可以的话，那么它是否能像普通私有函数一样访问私有属性？

（我不只是想试一下看会发生什么，因为我的项目现在有几个其他错误，而且我认为这是一个很好的参考问题。对我来说很难找到使用CUDA和C++的参考资料。可以找到基本功能示例，但没有结构化代码的策略。）

- George Aprilis

---

如果您真的想将内核与类关联起来，请考虑将指向__global__函数的指针作为（静态或常规）数据成员保留，或者使用getter函数返回它。 - einpoklum

Answer 1

暂时不考虑cuda动态并行性（即假设计算能力为3.0或更早版本）。

请记住，__global__用于cuda函数，这些函数将（仅）从主机调用（但在设备上执行）。如果您在设备上实例化此对象，则无法正常工作。此外，要使设备可访问的私有数据对成员函数可用，必须在设备上实例化该对象。

因此，您可以在主机对象成员函数中嵌入核心调用（即mykernel<<<blocks,threads>>>(...);），但是核心定义（即带有__ global__修饰符的函数定义）通常会在源代码中先于对象定义之前。如前所述，这种方法不能用于在设备上实例化的对象。它也无法访问在对象其他位置定义的普通私有数据。（可能可以想出一种方案，用于创建设备数据的仅主机对象，使用全局内存中的指针，然后在设备上访问该数据，但是这种方案在我第一眼看来似乎相当复杂）。

通常，可用于设备的成员函数应该以__device__修饰符为前缀。在这种情况下，设备成员函数中的所有代码都从调用它的线程中执行。

这个问题提供了一个示例（在我的编辑答案中），展示了一个C++对象，其成员函数可从主机和设备中调用，并且适当地复制主机和设备对象之间的数据。

- Robert Crovella