GPUプログラミング

上のようなループ(グリッドストライドループ)でカーネルを書く利点として正しいものはどれか。

for (int i = blockIdx.x * blockDim.x + threadIdx.x;
     i < n;
     i += blockDim.x * gridDim.x)
    y[i] = a * x[i] + y[i];
  1. Aループを使うと、メモリ合体アクセスにならなくなる
  2. B起動するスレッドの数が要素の数より少なくても、各スレッドが全スレッド数おきに複数の要素を処理して、すべての要素を扱える
  3. Cスレッドが1本しか動かなくなる
答えと解説を見る

正解:B. 起動するスレッドの数が要素の数より少なくても、各スレッドが全スレッド数おきに複数の要素を処理して、すべての要素を扱える

全スレッドの数(blockDim.x×gridDim.x)だけ番号を進めながら繰り返すので、スレッドの数がいくつでも全要素を漏れなく処理できます。GPUのSMの数に合わせたほどよい数のブロックで起動でき、スレッドの作成の手間も減ります。各回のループでは隣り合うスレッドが隣り合う要素を扱うので、メモリ合体アクセスも保たれます。

  • A. 各回のループで隣のスレッドは隣の要素を読むので、合体アクセスになります。
  • B. 正しい。グリッドの大きさを要素の数と切り離せます。
  • C. すべてのスレッドが動きます。
この分野を解く 分野ページへ

GPUプログラミングのほかの問題

  1. CUDAの定数メモリ(コンスタントメモリ)が特に速く働くのはどのようなアクセスのときか。
  2. GPUのキャッシュについて、CPUと比べた一般的な特徴として正しいものはどれか。
  3. NVIDIA以外のGPUでも使える、GPUなどの並列計算のためのプログラミングの仕組みの例として正しいものはどれか。
  4. GPUで、条件に合う要素だけを配列から取り出して別の配列に詰めて書き出したい(各スレッドが出力する数は0か1)。各スレッドの書き込み先の番地を決めるのによく使われる並列の計算はどれか。
  5. GPUのカーネルを書くときに守るべき前提として正しいものはどれか。