自分用メモ

https://wita.noyatsu.club/

CUDAメモ

CUDAでわからなかったもののリストです。 (マジで初歩)

用語

用語 意味
ホスト CPU側、プログラム側
バイス GPU
カーネル バイスで実行されるプログラムのこと(関数、メソッド的な)
スレッド カーネルを実行する最小単位のプロセスのこと, 各々のスレッドは3次元で管理される
ブロック スレッドのかたまり
グリッド ブロックのかたまり

カーネル

  • ホストメモリにはアクセス不可
  • 戻り値の型はvoid
  • 再帰処理はできない
  • 静的変数は使用不可
  • カーネルの前には__global__という関数識別子をつける

呼び出し方

カーネル名<<< dim3 グリッドに含まれるブロックの数(bNum), dim3 ブロックに含まれるスレッドの数(sNum)>>>(カーネルの引数)

※bNum×sNumの数が並列演算する回数(というか総スレッド数)になる

カーネル内で使える変数

__global__device__カーネルからアクセスできる。

変数 意味
blockIdx ブロックのインデックス
threadIdx ブロック内でのスレッドのインデックス
gridDim グリッドの次元(Max: 2D)
blockDim ブロックの次元
  • 3次元を扱うグリッド, ブロックの数やブロック, スレッドのインデックスは dim3型 で表現される
blockIdx.x * blockDim.x + threadIdx.x

を使うと一意のスレッドIDを生成可能。配列の添え字などに使用する。

使い方の例(配列の中身をすべてインクリメントするとき)

C++:

void inc_cpu(int *a, int N)
{
  int idx;
  for (idx = 0; idx<N; idx++)
    a[idx] = a[idx] + 1;
}

void main()
{
  ...
  inc_cpu(a, N);
}

CUDA:

__global__
void inc_gpu(int *a_d, int N)
{
  int idx = blockIdx.x * blockDim.x + threadIdx.x;
  if (idx < N)
    a_d[idx] = a_d[idx] + 1;
}

void main()
{
  …
  dim3 dimBlock (blocksize);
  dim3 dimGrid( ceil( N / (float)blocksize) );
  inc_gpu<<<dimGrid, dimBlock>>>(a_d, N);
}

メソッドについて

  • __mul24(x, y) …24bit以下の2つの整数の掛け算。x*yの結果(32bitまで)を返す。
  • ceil(x) ...xより大きい最小の整数を返す。