使用cudaMemcpy传输含const成员的Array结构体至CUDA设备的合法性疑问

首先给出你定义的Array类：

struct Array {
     float* const ptr;
     const size_t length;

     Array(float* const ptr, const size_t length) : ptr(ptr), length(length) {}
};

你在主机端通过placement new构造数组，再用cudaMemcpy拷贝到设备端的代码如下：

Array* arrays = (Array*) malloc(sizeof(Array) * 3);
new (arrays + 0) Array(nullptr, 0);
new (arrays + 1) Array(nullptr, 0);
new (arrays + 2) Array(nullptr, 0);

Array* device_arrays;
cudaMalloc(&device_arrays, sizeof(Array) * 3);

cudaMemcpy((void*) device_arrays, (void*) arrays, sizeof(Array) * 3, cudaMemcpyHostToDevice); 

核心问题解答 #

直接在核函数中使用device_arrays是合法且安全的，不需要在设备端重新构造Array对象。

原因在于：你的Array类属于可平凡复制（trivially copyable）类型——虽然它有用户定义的构造函数，但这个构造函数仅对const成员进行简单赋值，没有涉及非平凡操作（比如虚函数表初始化、动态内存分配、复杂对象依赖）。C++标准中，可平凡复制类型的内存内容可以直接拷贝，不需要调用构造/析构函数就能保证对象的有效性。

CUDA的cudaMemcpy本质是内存块的直接拷贝，对于可平凡复制类型，拷贝后的设备端内存内容和主机端完全一致，const成员的取值已经正确同步到设备端，核函数中直接访问device_arrays[0].length或ptr是完全合法的。

额外说明 #

如果你的类包含非平凡构造逻辑（比如带有虚函数、需要在构造时分配内存），那确实需要在设备端显式构造对象。但对于当前的Array类，直接内存拷贝的方式完全可行，不需要额外的POD中转或模板自动化操作。

内容的提问来源于stack exchange，提问作者Snowybluesky