1

它是否合并，如果n < warpSize？

// In kernel
int x;
if (threadId < n)
  x = globalMem[threadId];

这种情况出现在循环的最后一次迭代中，如果一些N不能被warpSize. 我应该在这些情况下运行并分配设备内存只能被整除warpSize还是按原样合并？

4

1 回答 1

1

如果按照cuda 编程指南 - 线程层次threadId结构中的记录正确计算，则此访问将被合并 - 这将是.threadId = threadIdx.x

对于不同的计算架构，内存合并略有不同。更多细节可以在cuda 编程指南的附录 G 中找到。

一般来说，如果您的线程从您的第一个线程访问的元素的地址开始抓取内存中的连续元素，则可以说全局内存访问是合并的。

假设您有一个浮点数组。
float array[]
你的记忆访问看起来是这样的

array[threadIdx.x == 0, threadId.x == 1, threadIdx.x == 2, ..., threadIdx.x == 31]

您的访问权限将被合并。

但是如果你以这种方式访问​​内存（例如交错）

array[threadIdx.x == 0, NONE, threadId.x == 1, NONE, threadIdx.x == 2, ..., NONE,  threadIdx.x == 31]

比你的访问没有合并（NONE意味着这个数组元素没有被任何线程访问）

在第一种情况下，您获取 128 个连续字节的内存。在第二种情况下，您获取 256 个字节。对于第二种情况，需要两个扭曲来从全局内存中加载内存，而不是第一种情况下的一个扭曲。但在这两种情况下，以下计算只需要 32 个浮点元素（即 128 个字节）。因此，在这种简单情况下，您的全局加载率将从 1.0 下降到 0.5。

于 2014-02-21T12:28:38.480 回答