memory - CUDA 合并和全局内存

Question

我在 CUDA 课程中被告知，如果我的“a”数组的元素大小为 4,8 或 16 字节，则以下访问（全局内存）将被合并。

int i = blockIdx.x*blockDim.x + threadIdx.x;
a[i];

合并的两个条件是：warp 的线程必须访问 32、64 或 128 字节的块。Warp 的第一个线程必须访问一个 32、64 或 128 倍数的地址

但是在这个例子中（第一个条件），没有任何东西可以保证 warp 会访问一个 32 字节的块。

如果我假设a的元素是浮点数（4字节），并且如果我将blockDim.x定义为5，那么即使我的“a”数组的元素大小为4，每个warp都将访问20（4x5）字节的块， 8 或 16 个字节，而不是 32 个字节。那么，关于合并的第一个声明是 false 吗？

谢谢您的回答。

score 2 · Accepted Answer

但是在这个例子中（第一个条件），没有任何东西可以保证 warp 会访问一个 32 字节的块。

由于线程排序，它保证每个 warp 访问 128 个字节（32 个线程 x 4 个字节）。这是合并内存访问的必要条件。

如果我假设a的元素是浮点数（4字节），并且如果我将blockDim.x定义为5，那么即使我的“a”数组的元素大小为4，每个warp都将访问20（4x5）字节的块， 8 或 16 个字节，而不是 32 个。

Warp 总是 32 个线程。如果将 blockDim.x 定义为 5，则每个块将包含 1 个带有 27 个空线程的 warp。合并规则仍将适用并且交易将被合并，但您正在浪费 27/32 的潜在计算能力和内存带宽。

那么，关于合并的第一个说法是错误的吗？

不。

1 回答 1