素人によるCUDAのお勉強。4

同期について

cudaでカーネル関数を実行している間は、同期を取らない限り、グリッド上で実行されているスレッドが終了しているとは限りません。明示的に同期を取るならば以下の関数があります。

cudaError_t cudaDeviceSynchronize(void);

一方で、暗に同期を取っている間すも存在し、cudaMemcpyもその内の一つです。これは、デバイスーホスト間のデータのコピーが終わるまで、処理が先に進むこがないことを保証しています。

cudaError_t cudaMemcpy(void* dst, const void* src, size_t count, cudaMemcpyKind kind);

カーネル関数

指示句

指示句	実行	注意
__global__	デバイス上で実行。ホストから呼び出し可能。	デバイスからの呼び出し出しは、capability3から可能。
__device__	デバイス上で実行。デバイスのみから呼び出し可能。
__host__	ホスト上で実行。ホストのみから呼び出し可能。

カーネル関数の制限

- デバイスのメモリにのみアクセス可能
- voidを返さなくてはならない。
- 可変長引数はサポートされていない。
- static変数はサポートされていない。
- 関数ポインターはサポートされていない。
- 非同時性の振る舞いをする→必要に応じて同期を取らないといけない。

エラーのハンドリング

#define CHECK(call) \
{ \
    const cudaError_t error = call; \
    if (error != cudaSuccess) \
    { \
        printf("Error: %s:%d, ", __FILE__, __LINE__); \
        printf("code:%d, reason: %s\n", error, cudaGetErrorString(error)); \
        exit(1); \
    }\
}

使い方。

CHECK(cudaMemcpy(d_C, gpuRef, nBytes, cudaMemcpyHostToDevice));

oimokihujin.hatenablog.com

Professional CUDA C Programming