ひびろぐ ver.2

Google検索したときの見た目を考慮してブログ名を変更してみた。馬鹿かと阿呆かと。

Archive for 8月 1st, 2007


CUDAの挙動を追いかけ中:共有メモリマルチスレッドプログラミングの実験みたいなもの

共有メモリに同時多数アクセスをしたらどうなるのか。

CUDA:

  1. __global__ void test
  2. (
  3.  int size
  4.  , int* nums
  5.  )
  6. {
  7.   int id = threadIdx.x + threadIdx.y*10;
  8.   __shared__ int data[100];
  9.   data[0] = id;
  10.   nums[id] = data[0];
  11. }


適当なサイズの共有メモリを確保して、いわゆるスレッド番号みたいなものを書き込んで、取得してみる。

同一の共有メモリにみんなで書き込んでいるため、値がどうなるかわかったもんじゃない。

共有メモリマルチスレッドプログラミングのバグの温床。

のはずなんだけど、値が狂わない。おかしいなあ?


CPUプログラミングで言うとことのスレッドであるCUDAのブロックは互いに同期を取れないらしいんだけど、これって一般的な並列プログラミングの足枷になるよなあ……。協調処理がしにくすぎる。アプリケーションとの相性になるよなぁ。

英語が分かるかとかいう問題ではない

何故か先日(7/26)投稿したはずの記事が草稿扱いだったので、再投稿。

以下本文。


東大構内で外人に道を聞かれた。


マヂごめんなさい。

東大の人じゃなくてごめんなさい。

……目的地、ここ(小規模国際ワークショップ開催中)じゃないよね?

CUDAの並列実行モデル?についてのメモ

自分のための備忘録。というか自己解釈。間違ってる部分もあるかも。

GPUへ投げる``ジョブ''の単位は``Grid''。


``Grid''は複数の``Block''で構成される。

1つの``Block''は1つの``Multiprocessor''上で実行される。

(``Block''が更に分割されて``Multiprocessor''に割り当てられることは無い。)

既存のCPUプログラミングにおけるスレッド的な位置づけだが、``Block''間の同期は取れない。

``Multiprocessor''の数が並列実行可能な``Block''の数に対応。

``Block''の数が``Multiprocessor''の数を超えた場合は、

適当な``Multiprocessor''に複数の``Block''が割り当てられ、

タイムスライシングで平行実行される。


``Block''は複数の``Thread''で構成される。

``Thread''は既存のCPUプログラミングで言うところのSIMD的な位置づけ。

感覚的にはむしろ、OpenMPのループ並列に近く感じられる。

``Block''内で同時実行可能な``Thread''の数は8(=Multiprocessorを構成するプロセッサの数)。

warp size=32はどこに関係するんだろう?



うーむ、HWのモデルとメモリのモデルとがうまく脳内にマッピングしきれない。

とりあえず実行時間ベースでの調査に進むか。