既存コードでCUDAを使う

既存コードにCUDAのコードを追加する際にはまったためのメモ

VisualStudioは持っていたが、C#しか使用していなかったので、既存コードなんてないのですが、
新規プロジェクトを生成してから、CUDAを使おうとして嵌った。

1. VCのソリューションエクスプローラーで、プロジェクトを右クリック、ビルドのカスタマイズを選択し、使用できるビルドカスタマイズファイルの "CUDA 4.0 (targets, props)"にチェックを入れる。
   このままだと
2. プロジェクトにCudaBuildTasksPathが設定されていないためなので、.vcxprojをテキストファイルで開き、下記の様に修正する。 ( Blogger 作成とHTMLを切り替えるとおかしくなるため行の先頭と末尾の　< > は省略した。　）
   

   修正前	Import Project="$(VCTargetsPath)\Microsoft.Cpp.props" / ImportGroup Label="ExtensionSettings" /ImportGroup
   修正後	Import Project="$(VCTargetsPath)\Microsoft.Cpp.props" / ImportGroup Label="ExtensionSettings"  Import Project="$(VCTargetsPath)\BuildCustomizations\CUDA 4.0.props" / /ImportGroup

3. ソリューションエクスプローラーでプロジェクトを右クリックしプロパティを選択、"リンカー-入力"を選択、追加の依存ファイルに、cudart.dllを追加する。
4. .cuファイルを追加
   ソリューションエクスプローラーで、.cuファイルのプロパティを開けて、項目の種類をCUDA C/C++に変更する。

http://developer.nvidia.com/cuda-toolkit-40　に　*NEW* CUDA Toolkit 4.0 Build Customization BUG FIX Update Fixes  error message "$(CUDABuildTasksPath) property is not valid" という項目があったが、適応していなかったため嵌っていたようです。

CUDA プロジェクトの作成

CUDA_SDK_Release_Notes.txtよりCUDAのプロジェクトの作り方を適当に訳した

1. "NVIDIA GPU Computing SDK 4.0\C\src"にある"template"か"template_runtime"の内容を、"NVIDIA GPU Computing SDK 4.0\C\src\myproject"にコピーする。
2. プロジェクトのファイル名を好みに合わせて変更する。
3. "*.sln","*.vcproj"およびソースファイルを変更する。
   変更内容: "template"または"template_runtime" -> "myproject"
4. "myproject_vs2005.sln","myproject_vs2008.sln"または"myproject_vs2010.sln"を使ってビルドを行なう。
5. "NVIDIA GPU Computing SDK 4.0\C\bin\win[32|64]\[release|debug]"にある"myproject.exe"を実行。
6. 計算に必要なコードを編集する。

* ｔemplate と template_runtime の大きな違い：template は、cutilを使用しているが、 template_runtimeは使用していない。

cutilは便利なマクロが用意されているが、標準エラー出力を使うため、独自のエラー処理が必要な場合は、template_runtimeを使用した方がよい。

CUDA API #3 ディバイス制御

• cudaGetDeviceCount(int*)
  システムに実装されているCUDAのディバイス数の取得
• cudaGetDeviceProperties(cudaDeviceProp*,int)
  CUDAディバイスのプロパティ取得
• cudaSetDevice(int)
  アクティブなCUDAディバイスの設定
• cudaGetDevice(int*)
  アクティブなCUDAディバイスの取得

• cudaDeviceReset()
  ディバイスリセット　（CUDA4.0以上)
• cudaThreadExit()
  スレッド終了　（CUDA4.0未満では、ディバイスリセットとして使用されている)
• cudaDeviceSynchronize()
  実行の同期(CPU←→GPU間?)
  カーネルの実行は非同期で行なわれるため、CPUに結果を転送する場合はこの関数でカーネル処理利が終了するまで待つ必要がある。
• cudaThreadSynchronize()
  実行の同期　　（CUDA4.0未満)

CUDA API #2 エラー処理

CUDAのサンプルでは、cutilを用いてエラー処理していますが、NVIDIA CUDA C GETTING STARTED GUIDE FORMICROSOFT WINDOWSには下記の記述があるため。

元	The cutil library is primarily intended as a tool for streamlining the GPU Computing SDK samples; its use outside of the SDK samples is not recommended, as the API for the library may change at any time.
訳	cutilライブラリは主にGPUコンピューティングSDKのサンプルを合理化するためのツールとして意図され、ライブラリのAPIは、いつでも変更される可能性があるとして、SDKサンプルの外で、その使用は、推奨されません.

と記述があるため。cutilを使用しない、エラーチェックにつて記録しておく。


cudaError_t ：　多くのcuda関数は戻り値型
カーネル呼び出しは戻り値がvoidであるため、cudaGetLastErrorで最後のエラーコード取得する。
char* cudaGetErrorString(cudaError_t)でエラーメッセージを取得できる。

CUDA カーネル

カーネル側のプログラム
__global__修飾して宣言、戻り値はvoid固定で下記の様に実装する。
    __global__ void cudaFunction(inputs,outputs) { ... }

カーネルの呼び出しは下記の様に<<< >>>の内側に、ブロック数、スレッド数を指定する拡張表記を使用する。
    cudaFunction<<< blocks, threads >>>( input, output );

計算単位での値の設定取得は、ポインタなどを用いる。
ポインタの参照位置は下記の定義済み変数を用いて算出する。
一般的なアドレス算出は下記の用になる。
int index = threadIdx.x
+ threadIdx.y * blockDim.x
+ blockIdx.x * blockDim.x * blockDim.y
+ blockIdx.y * blockDim.x * blockDim.y * gridDim.x
+ blockIdx.z * blockDim.x * blockDim.y * gridDim.x * gridDim.y;

カーネルで使用可能な定義済み変数一覧

変数名	型	内容
gridDim	dim3	gridのサイズ
blockIdx	uint3	grid内のブロックの座標
blockDim	dim3	blockのサイズ
threadIdx	uint3	block内のスレッドの座標

dim3,int3とも下記の構造体であるが、dim3は各メンバとも初期値が1となる。

struct {
    int x;
    int y;
    int z;
};

CUDA API #1 メモリ管理

• cudaMalloc(void **devptr, size_t count)
  デバイスメモリ確保。
• cudaFree(void* devptr)
  ディバイスメモリ開放
• cudaHostMalloc(void **hostptr,size_t count)
  ホストメモリ確保。
  ページロックされたメモリが確保できるためmallocで確保したメモリより高速にアクセスできる可能性が高い
• cudaHostFree(void* hostptr)
  ホストメモリ開放
• cudaMemcpy(void *dst, const void *src, size_t count, enum cudaMemcpyKind kind)
  メモリコピー
  

  cudaMemcpyKindの値

  cudaMemcpyHostToDevice	dstはディバイスメモリ、srcはホストメモリ
  cudaMemcpyDeviceToHost	dstはホストメモリ、srcはディバイスメモリ
  cudaMemcpyDeviceToDevice	dst,srcともディバイスメモリ

CUDA インストール

CUDAのインストールを忘れないようにメモ
NVIDIA CUDA C GETTING STARTED GUIDE FORMICROSOFT WINDOWSを参照しながら実施。

1. ダウンロード
   Windows7（64bit）でVisualStudio2010を使うので、ダウンロードページから　cudatoolkit、gpucomputingsdkの64bit版をダウンロード。ドライバはアップされているものより新しいものを使用しているのでパスした。
2. インストール
   cudatoolkit、gpucomputingsdkを順にインストール
   環境変数も設定される。
3. インストールの確認
1. コマンドプロンプトで、nvcc -vを実行。　実行されてエラーが出ればOK
2. bandwidthTesｔの実行。　test resultsがPASSEDと出ればOK
   ("C:\ProgramData\NVIDIA Corporation\NVIDIA GPU Computing SDK 4.0\C\bin\win64\Release")
4. simpleD3D9のビルド
   そのままビルドすると、cutilが見つからずビルドエラーになる。
1. "C:\ProgramData\NVIDIA Corporation\NVIDIA GPU Computing SDK 4.0\C\common"にあるある"cutil_vs2010.sln" を開きすべてバッチビルド
2. simpleD3D9のビルド
   ここでもビルドエラーなら、DirectX SDKがインストールされていないとか…
5. 追加設定
   ドキュメントには、"$VisualStudioInstallDir\VC\VCProjectDefaults"に.rulesファイルがインストールしてあるように記述されているが、コピーされていないため、下記の操作を実施しておく。
1. "C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v4.0\extras\visual_studio_integration\rules"のファイルを"C:\Program Files (x86)\Microsoft Visual Studio 10.0\VC\VCProjectDefaults"にコピー。
6. その他
1. バグフィックスのダウンロード
   ダウンロードページから"*NEW* CUDA Toolkit 4.0 Build Customization BUG FIX Update Fixes  error message "$(CUDABuildTasksPath) property is not valid""をダウンロード。
2. README_CUDA4-0-BuildCustomization.txtの指示に従い
   32 bit OS:  "C:\Program Files\MSBuild\Microsoft.Cpp\v4.0\BuildCustomizations"
   64 bit OS:  "C:\Program Files (x86)\MSBuild\Microsoft.Cpp\v4.0\BuildCustomizations"に　"CUDA 4.0.targets"と"CUDA 4.0.props"をリプレイス

note:翻訳メモ

The cutil library is primarily intended as a tool for streamlining the GPU Computing SDK samples; its use outside of the SDK samples is not recommended, as the API for the library may change at any time.	cutilライブラリは主にGPUコンピューティングSDKのサンプルを合理化するためのツールとして意図され、ライブラリのAPIは、いつでも変更される可能性があるとして、SDKサンプルの外で、その使用は、推奨されません.
These sample projects also make use of the $CUDA_PATH environment variable to locate the CUDA Toolkit and a .rules file to locate and configure the nvcc compiler.	サンプルプロジェクトは、CUDAツールキットの位置を占めす$CUDA_PATH環境変数と、nvccコンパイラの位置と設定を示す.rulesファイルを使用します。
The environment variable is set automatically and the .rules file is installed automatically as part of the CUDA Toolkit installation process.	環境変数が自動的に設定され、.rulesファイルは、CUDA Toolkitのインストールプロセスの一部として自動的にインストールされます。
The .rules file is installed into $VisualStudioInstallDir\VC\VCProjectDefaults.	.rulesファイルは、$VisualStudioInstallDir\VC\VCProjectDefaultsにインストールされています。
You can reference this .rules file from your Visual Studio project files when building your own CUDA applications.	あなた自身のCUDAアプリケーションを構築する際には、Visual Studioのプロジェクトファイルから.rulesファイルを参照することができます。