我打开这个主题是因为我注意到我的代码输出中有奇怪的行为，同时试图深入了解 CUDA 中的一些基本概念，如速度与块/线程数等......任何帮助将不胜感激！
首先，这是我显卡的一些规格：
名称：GeForce 8600M GT
多处理器数量：4
每个块的最大线程数：512
最大网格尺寸：(65535, 65535, 1)
我正在玩以下简单的代码。它用 1 填充 3 个长度为 N 的数组并计算总和。总和显然是可预测的，等于 3N。
#include <iostream>
#include "ArgumentParser.h"
//using namespace std;
__global__ void addVector(int *a, int *b, int *c, int *d, int *N){
int tid = threadIdx.x + blockIdx.x * blockDim.x;
if (tid<*N) {
d[tid] = a[tid] + b[tid] + c[tid];
}
}
int main(int argc, char *argv[]) {
//Handy way to pass command-line arguments.
ArgumentParser parser(argc, argv);
int nblocks = parser("-nblocks").asInt(1);
int nthreads = parser("-nthreads").asInt(1);
//Defining arrays on host.
int N = 100000;
int a[N];
int b[N];
int c[N];
int d[N];
//Pointers to the arrays that will go to the device.
int *dev_a;
int *dev_b;
int *dev_c;
int *dev_d;
int *dev_N;
//Filling up a, b, and c.
for (int i=0; i<N; i++){
a[i] = 1;
b[i] = 1;
c[i] = 1;
}
//Modifying the memory adress of dev_x so that dev_x is on the device and //
//the proper memory size is reserved for it.
cudaMalloc((void**)&dev_a, N * sizeof(int));
cudaMalloc((void**)&dev_b, N * sizeof(int));
cudaMalloc((void**)&dev_c, N * sizeof(int));
cudaMalloc((void**)&dev_d, N * sizeof(int));
cudaMalloc((void**)&dev_N, sizeof(int));
//Copying the content of a/b/c and N to from the host to the device.
cudaMemcpy(dev_a, a, N * sizeof(int), cudaMemcpyHostToDevice);
cudaMemcpy(dev_b, b, N * sizeof(int), cudaMemcpyHostToDevice);
cudaMemcpy(dev_c, c, N * sizeof(int), cudaMemcpyHostToDevice);
cudaMemcpy(dev_N, &N, sizeof(int), cudaMemcpyHostToDevice);
//Initializing the cuda timers.
cudaEvent_t start, stop;
cudaEventCreate(&start);
cudaEventCreate(&stop);
cudaEventRecord (start, 0);
//Executing the kernel.
addVector<<<nblocks, nthreads>>>(dev_a, dev_b, dev_c, dev_d, dev_N);
cudaEventRecord(stop, 0);
cudaEventSynchronize(stop);
float time;
cudaEventElapsedTime(&time, start, stop);
printf ("CUDA time: %3.5f s\n", time/1000);
//Copying the result from device to host.
cudaMemcpy(d, dev_d, N * sizeof(int), cudaMemcpyDeviceToHost);
//Freeing the memory allocated on the GPU
cudaFree(dev_a);
cudaFree(dev_b);
cudaFree(dev_c);
cudaFree(dev_d);
cudaFree(dev_N);
//Checking the predictable result.
int sum=0;
for (int i=0; i<N; i++){
sum += d[i];
}
printf("Result of the sum: %d. It should be: %d.\n", sum, 3*N);
}
问题1：
当我编译代码并输入时：
./addArrayCuda -nblocks 1 -nthreads 1
我得到一个答案：
Result of the sum: -642264408. It should be: 300000.
这似乎是合理的。我使用带有单个线程的单个块。只会添加每个数组的第一个元素。其余元素是一些随机值，它们加起来是不可预测的。应该是 nblocks * nthreads >= N。所以让我们试试：
./addArrayCuda -nblocks 3125 -nthreads 32
输出是：
Result of the sum: 300000. It should be: 300000.
这是有道理的。3125 * 32 = 100000 = N。直到这里，一切都很好。但是，如果我重新运行上一个命令（使用 nblocks = nthreads = 1）而不重新编译，我会得到：
./addArrayCuda -nblocks 1 -nthreads 1
Result of the sum: 300000. It should be: 300000.
发生了什么？？
问题 2： 这个问题是关于 nblocks/nthreads 与执行速度之间的关系。我知道如果代码中的问题解释了Question 1 ，这个问题可能没有太大意义，但我还是问一下吧。我已经查看了具有不同数量的块/线程的代码的执行时间（平均超过 5 次运行），但要确保 nblocks * nthreads > N。这就是我所得到的（我有一个很好的情节但不是足够的声誉来发布它......）：
(nblocks, nthreads) 执行时间 [s] 增加比率
(196, 512) 5.0e-4 -
(391, 256) 4.8e-4 1.0
(782, 128) 4.8e-4 1.0
(1563, 64) 4.9e- 4 1.0
(3125, 32) 5.0e-4 1.0
(6250, 16) 5.2e-4 1.0
(12500, 8) 9.0e-4 1.7
(25000, 4) 1.3e-3 1.4
(50000, 2) 2.3e- 3 1.8
我的解释：GPU被分成块，每个块被分成线程。GPU 的每个时钟周期将内核发送到 4 个块（多处理器计数），并在每个块内发送到一个 warp（32 个线程组）。这意味着使用不是 32 倍数的线程数是对资源的浪费。因此，我们可以理解（nblocks，nthreads）与执行时间之间的一般关系。从 (196, 512) 到 (3125, 32)，GPU 占用的时钟周期数大致相同，大约与 (nblocks / 4) * (nthreads / 32) 成正比。但是，我们粗略预计 (3125, 32) 和 (6250, 16)、(6250, 16) 和 (12500, 8) 等之间的执行时间会增加一倍。
为什么不是这样？更具体地说，为什么 (3125, 32) 和 (6250, 16) 之间的执行时间没有显着差异？
感谢您花时间阅读到这里 ;-)