このトピックを開いているのは、速度とブロック/スレッドの数などの CUDA のいくつかの基本的な概念について洞察を得ようとしているときに、コードの出力に奇妙な動作があることに気付いたからです...助けていただければ幸いです!
まず最初に、私のグラフィック カードの仕様をいくつ
か
示し
ます
。
私は次の簡単なコードで遊んでいました。長さ N の 3 つの配列を 1 で埋め、合計を計算します。合計は明らかに予測可能で、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 と実行速度の関係に関するものです。コードの問題が質問 1を説明している場合、この質問があまり意味をなさない可能性があることは承知していますが、それでも質問させてください。さまざまな数のブロック/スレッドでコードの実行時間 (5 回の実行の平均) を調べましたが、nblocks * nthreads > N であることを確認しました。それを投稿するのに十分な評判...）：
(nblocks, nthreads) 実行時間 [秒] 増加率
(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 つのブロック (マルチプロセッサ カウント) に送信し、それらの各ブロック内でワープ (32 スレッドのグループ) に送信します。これは、32 の倍数ではない数のスレッドを使用すると、リソースが無駄になることを意味します。そのため、(nblocks、nthreads) と実行時間の間の一般的な関係を理解できます。(196, 512) から (3125, 32) まで、GPU が使用するクロック サイクル数はほぼ同じで、(nblocks / 4) * (nthreads / 32) にほぼ比例します。ただし、(3125, 32) と (6250, 16)、(6250, 16) と (12500, 8) などの間では、実行時間が 2 倍になると大まかに予想されます。
そうでないのはなぜですか？より具体的には、(3125, 32) と (6250, 16) の間で実行時間に有意差がないのはなぜですか?
ここまで読んでくれてありがとう(;_;)