CUDA 分配对齐是 256 字节 - 认真吗？

Question

在“CUDA C 编程指南 5.0”中，p73 (also here) 说“驻留在全局内存中或由驱动程序或运行时 API 的内存分配例程之一返回的变量的任何地址始终与至少 256 个字节对齐” .我不知道这句话的确切含义。谁能给我举个例子？非常感谢。

一个衍生问题:
那么，如何分配基本元素(如 int)或自定义元素的一维数组呢？数组的起始地址会是256B的倍数，而数组中每个元素的地址不一定是256B的倍数？

Answer 1

使用 CUDA 运行时的任何设备内存分配函数分配的指针，例如 cudaMalloc或 cudaMallocPitch保证是 256 字节对齐的，即地址是 256 的倍数。

考虑以下示例:

char *ptr1, *ptr2;

int bytes = 1;

cudaMalloc((void**)&ptr1,bytes);
cudaMalloc((void**)&ptr2,bytes);

假设 ptr1 中返回的地址是 256 的某个倍数，则返回 ptr2 中的地址至少会是 (ptr1 + 256) .

这是分配内存的设备所施加的限制。大多数情况下，指针是出于性能目的而对齐的。 (一些 NVIDIA 的家伙应该能够判断是否还有其他原因)。

重要的:

指针对齐并不总是 256 .在我的设备 (GTX460M) 上，它是 512。您可以通过 cudaDeviceProp::textureAlignment 获得设备指针对齐 field 。

指针的对齐也是将指针绑定(bind)到纹理的要求。