c++ - 如何将 cudaMallocManaged 与带有构造函数的对象一起使用

Question

我正在尝试将 CUDA 托管内存与我通过构造函数创建的对象结合使用。

struct A {
    A(float x) : x(x) {}
    float x;
}

__global__ void myKernel(A *a) {
    printf("%f", a->x);
}

int main() {
    A *a;
    cudaMallocManaged(&a, sizeof(A));
    a->x = 42.f;        // This works
    // a = new a(42.f); // This obviously doesn't because a doesn't point
                        // to managed memory now.
    myKernel<<<1,1>>>(a);
}

使用构造函数而不是直接初始化 a->x = ... 对于 A 更复杂的类来说会很方便。

我当然可以使用构造函数创建对象 a 并使用“普通”cudaMalloc 和 cudaMemcpy，如下所示。

A a(42.f);
A *d_a;
cudaMalloc(&d_a, sizeof(A));
cudaMemcpy(d_a, &a, sizeof(A), cudaMemcpyHostToDevice);

是否有可能使用托管内存并仍然使用它们的构造函数初始化对象？

Answer 1

正如 Mark Harris 所讨论的那样 here ，您可以为您的类重载 new 和 delete 运算符，以便通过托管内存 API 分配和释放实例，而不是通过主机标准库内存分配。这样做意味着您不能在设备代码中为您以这种方式定义的任何类调用 new。

一个非常方便的设计模式(这个想法再次完全归功于 Mark Harris 博客)是定义一个只包含合适的 new 和 delete 的类运算符，然后在您定义自己的将使用统一内存的类时继承它。像这样:

class Managed {
public:
  void *operator new(size_t len) {
    void *ptr;
    cudaMallocManaged(&ptr, len);
    cudaDeviceSynchronize();
    return ptr;
  }

  void operator delete(void *ptr) {
    cudaDeviceSynchronize();
    cudaFree(ptr);
  }
};

class A : public Managed
{
public:
    A(float x) : x(x) {}
    float x;
}

// ....

A *a = new A(42.f);

请注意，理想情况下，您还应该为数据类定义赋值运算符，以便复制构造和其他有用的 C++ 惯用语能够正常工作。