为什么我们在 CUDA 内核中需要 stride？

Question

我想知道，为什么需要在以下循环中使用grid-stride步幅：

for (int i = index; i < ITERATIONS; i =+ stride)
{
  C[i] = A[i] + B[i];
}

我们将步幅和索引设置为：

index = blockIdx.x * blockDim.x + threadIdx.x;
stride = blockDim.x * gridDim.x;

当调用内核时，我们有这个：

int blockSize = 5;
int ITERATIONS = 20;
int numBlocks = (ITERATIONS + blockSize - 1) / blockSize;
bench<<<numBlocks, blockSize>>>(A, B, C);

因此，当我们启动内核时，我们将拥有 blockDim.x = 5 和 gridDim = 4，并且步幅将等于 20。

我的观点是，每当使用这种方法时，步幅总是等于或大于计算中的元素数量，因此每次增量循环都会结束。

问题是，为什么需要使用循环或跨步，为什么不像这样使用索引运行？:

index = blockIdx.x * blockDim.x + threadIdx.x;
C[index] = A[index] + B[index];

还有一个问题，在这种特殊情况下，我现在如何在“跳转”到一个非常大的数组的另一部分（例如 2000000）之前同时在我的 GPU 上运行多少个线程？

Answer 1

我的观点是，每当一个人使用这种方法时，stride 总是会等于或大于计算中的元素数量，因此每个到了增量循环的时间就结束了。

你的理解有问题。为了有效地使用该内核，您只需要运行尽可能多的块来实现设备的最大设备范围占用，而不是运行处理所有数据所需的块。然后，这些较少的块将成为“常驻”，并为每个线程处理多个输入/输出对。网格步幅还保留内核可能具有的任何内存合并和缓存一致性属性。

通过这样做，您可以消除调度和退休块的开销。这样做可以在简单内核中显着提高效率。这种设计模式没有其他原因。

Answer 2