メニーコアと格闘ブログ:CUDAを使ってプログラムとC言語プログラムの比較 その2

出典: トータル・ディスクロージャ・サイト(事実をありのままに)

[プログラム解説]

ad.cについては、普通のC言語プログラムです。integration関数を呼び出して積分を行う、ということで動作は直観的に理解できると思います。


ここから先では、ad1.cuについて説明します。

メニーコアと格闘ブログ:CUDA 漫ろ歩き にも述べましたが、一般的なC言語プログラミングとCUDAでのプログラミングの違いとして、以下のことが挙げられます。

  • [CUDAアプリケーション] = [ホストで動作するコード] + [デバイスで動作するコード]
  • CUDAでは次のメモリ領域を明確に区別

  1.ホスト側メモリ

  2.デバイス側メインメモリ

  3.デバイス側テクスチャメモリ

  4.デバイス側コンスタントメモリ

  5.デバイス側共有メモリ

今回は、ホスト側メモリとデバイス側メモリについて、扱っています。



[順にプログラムの流れを追ってみましょう]

CUDAプログラミングでは(も)、まず、main関数が実行されます。

cudaMallocにて“デバイス側の”メモリが確保されます。ここでは u、du という名前の配列が確保される形になります。

cudaMallocHostにて、“CPU側の”メモリが確保されます。ここでは u_h という名前の配列が確保されました。

“デバイス側の” “CPU側の”と、わざわざ強調したのはなぜかというと、CUDAプログラミングでは、メモリ領域が独立管理されるからです。つまり、書きようによっては、デバイス側での変数名と、ホスト側の変数名が同一でも、構わない仕様になっています。

set_initialに処理が移り、cudaMemcpy()にてHOSTからDEVICEへデータがコピーされます。

integration_kernel0が呼び出されます。この関数は__global__修飾子で宣言されています。この修飾子が付された関数はGPUで動作する処理です。

integration_kernel1が呼ばれます。④と同じく、__global__修飾子で宣言されていますので、CPU側から呼ばれる関数として認識されます。

cudaMemcpy()として、指定した位置から別の場所へデータをコピーします。ここでは、“cudaMemcpyDeviceToDevice”を指定しているので、デバイス内でのデータコピーということになります。

CPU側の処理として、outputが呼ばれます。ここでは、データを指定のファイルへ書き出すことをしますが、GPUは直接ファイルを編集できないので、GPUからCPUへデータをコピーする必要があります。ここでは、cudaMemcpy()にて“cudaMemcpyDeviceToHost”を指定して、データコピーを行います。


ファイル:CUDAを使ってプログラムとC言語プログラムの比較その2-01.jpg


[カーネル関数の呼び出しについて]

④、⑤のように、CPU側から呼ぶ関数を“カーネル関数”と呼びます。

カーネル関数の呼び出し方は、

   kernel<<<dim3 dG, dim3 dB>>>(…)


各変数の意味は、以下のようになります。

  • kernel=カーネル関数名。
  • dG=グリッド数
  • dB=ブロック数
  • (カッコ)内は引数


ここで、グリッド数、ブロック数はソフト的な概念の呼称です。要するに、“グリッド数xブロック数分のスレッドが起動する”と、いう意味になります。 [参照:CUDAプログラミングモデルの概要 P11:[1]]

  • “グリッド数”とは、スレッドブロックの集まり。
  • “ブロック数”とは、スレッドの集まりの最小単位。


ここでの注意点として、全てのスレッドが同じカーネル関数を実行します。スレッドはそれぞれがIDを持ち、このIDによって振舞を変えます


例のプログラムで言うと、カーネル関数に入った直後の、

int n = BLOCKDIM * blockIdx.x + threadIdx.x + 1;

という部分が、振舞の分かれ目に当たります。このような操作は、MPI並列で言うところの、rankによる振舞の振り分けの概念と共通するところと言えます。

ナン 2009年8月20日 (木) 15:28 (JST)


この記事へのコメントをお寄せください

  • サイトへの書き込みに差し支えございましたら トータルディスクロージャーサイトサポート係へメールをお送りください
  • トータル・ディスクロージャ・サイトに投稿された文章と画像は、すべてその著作権がHPCシステムズ株式会社に帰属し、HPCシステムズ株式会社が著作権を所有することに同意してください。
  • あなたの文章が他人によって自由に編集、配布されることを望まない場合は、投稿を控えてください。
  • コメントを書き込む場合は名前にひらがなを織り交ぜてください。
  • あなたの投稿する文章と画像はあなた自身によって書かれたものであるか、パブリック・ドメインかそれに類する自由なリソースからの複製であることを約束してください。あなたが著作権を保持していない作品を許諾なしに投稿してはいけません!

<comments hideform="false" />


Comments

ノート:メニーコアと格闘ブログ:CUDAを使ってプログラムとC言語プログラムの比較 その2

個人用ツール