トップ > C++/CLIでCUDA

C++/CLIでCUDAを使うには

VisualC++ 2005の.NET Framework 2.0を用いて、CUDAを使うには、GASSのCUDA.NETを使うのがもっとも楽です。

CUDAの特徴と癖

GPGPU(CUDA)を使う上での大きな特徴は数百のプロセッサを制御し、多目的にプログラミングが可能なことであると思います。
しかしながら、その反面、独特な"癖"を持っています。どのように、その癖を把握しつつ、制御して目的を達成させるかがポイントであるといえます。

• メモリアクセスのプログラム記述で処理時間が一桁変わる
• 処理同士が情報を共有する場合には向かない
• 一つのプログラムを全てのプロセッサで動作させる（SIMD）ため、プロセッサごとに異なる処理をさせることはできない
• PCIexpressの帯域幅が狭いため、GPUへのデータ転送時間より計算処理時間が短い処理では効果が薄い

CUDA.NETの使用方法

予め、CUDA 2.x SDKとCUDA Tool Kit、ドライバをインストールしている事。Tool Kitが使えるように、Visual C++2005のbinのPATH(環境変数)も通しておくこと。
CUDA.NETはどこかのフォルダに解凍しておき、CUDA.NET.DLLを参照するようにプロジェクトのプロパティを設定すればよい。通常の.NETのDLLの参照を行う作業と同様です。

C++/CLI側からの使い方

・usingでの定義に２つのネームスペースを追加
using namespace GASS::CUDA;
using namespace GASS::CUDA::Types;
・関数内でのCUDA実行プログラム(cubin)のロードと実行まで
CUDA^ cuda;
CUfunction cudaFunc;
array^ pixels;//CUDAに渡す配列
array^ pixelsResult;//CUDAから受け取る配列格納先
int step = 128;//CUDAに渡す引数

//CUDA.NETの初期化と関数ロード
cuda = gcnew CUDA(0, true);
cuda->LoadModule("cubinファイル名");
cudaFunc = cuda->GetModuleFunction("関数名");

//デバイスに画像を転送
CUdeviceptr pixs = cuda->CopyHostToDevice(pixels);

//実行
cuda->SetParameter(cudaFunc, 0, (unsigned int)pixs.Pointer.ToInt32());
cuda->SetParameter(cudaFunc, 4, (unsigned int)step);
cuda->SetParameterSize(cudaFunc, 8);

//ブロックとグリッドの設定と実行
cuda->SetFunctionBlockShape(cudaFunc, 32, 1, 1);
cuda->Launch(cudaFunc, pixels->Length / 32, 1);

//処理の完了待ち
cuda->SynchronizeContext();

//結果データ取得
cuda->CopyDeviceToHost(pixs, pixelsResult);

プログラミングガイドやサンプルでは、CUDAの実行プログラムとホストPC側のアプリケーションは混在してプログラムをしています。
CUDA.NETでは、C++/CLIなのでそのようなことはできません。そこで、CUDAドライバのランタイムAPIを利用して、CUDAプログラムのみをコンパイルして作ったcubinファイルをホストPC側のアプリケーション実行時にロードし、転送して実行させます。

GPUで実行するCUDAプログラム例

extern "C" __global__ void filter(int *pixels, int step)
{
int idx = blockIdx.x * blockDim.x + threadIdx.x;

int value = pixels[idx];

int r = (value >> 16) & 0xFF;
int g = (value >> 8) & 0xFF;
int b = value & 0xFF;

r = (r + g + b) / 3;

pixels[idx] = (r + (r << 8) + (r << 16)) | 0xFF000000;
}

感想

GPGPUは演算処理に関してはCPUよりもずっと速いのは間違いありません。しかしながら、分岐命令・メモリアクセス（特にランダムアクセス）が非常に遅い。PCIeも遅すぎる。結局、CPUでやったほうが速い状況が多いというのが現在の結論です。
ただし、今後はCPUと直結されるという方向がメーカーから示されていますので、こうなれば状況は変わると思います。

ちなみにメモリアクセスが遅いのは、メモリコントローラに妙な制約が多い（コアレスアクセスの発動条件が酷い）からなので、これはNVIDIAにがんばってもらわないといけない。