というわけで再挑戦。
nvidiaのサンプルに倣って、gpuの処理を書いたgpukernel.cuをcuda.cuからincludeしてみた。
まずはCPUの処理を記述しているcuda.cu。
CU:
- #include <stdlib.h>
- #include <stdio.h>
- #include <string.h>
- #include <math.h>
- #include <cutil.h>
- // サンプルの真似をしてGPUの計算を一応別ファイル化
- #include "gpukernel.cu"
- void
- runTest(int argc, char** argv)
- {
- CUT_DEVICE_INIT();
- int nSize = 10;
- int n;
- n = sizeof(float)*nSize;
- // CPU側のデータを準備
- float *h_sd1;
- h_sd1 = (float*)malloc(n);
- int i;
- for(i=0; i<nSize; i++){
- h_sd1[i] = (float)(rand()%100)/10.0f;
- }
- // 値の確認
- for(i=0; i<nSize; i++){
- printf(" %f", h_sd1[i]);
- }
- printf("\n");
- // GPU側のメモリを準備
- float* d_sd1;
- CUDA_SAFE_CALL(cudaMalloc((void**)&d_sd1, n));
- CUDA_SAFE_CALL(cudaMemcpy(d_sd1, h_sd1, n, cudaMemcpyHostToDevice) );
- // 演算結果取得のためのメモリを準備
- float *h_sd2;
- h_sd2 = (float*)malloc(n);
- float* d_sd2;
- CUDA_SAFE_CALL(cudaMalloc((void**)&d_sd2, n));
- // 実行のためのパラメタを準備
- dim3 threads(nSize, 1, 1);
- dim3 grid(1,1,1);
- // 実行
- test<<<grid, threads>>>(d_sd2, d_sd1, nSize, 2.0f);
- // GPUの処理に問題が起きていないかの確認
- CUT_CHECK_ERROR("Kernel execution failed");
- // 演算結果の取得
- CUDA_SAFE_CALL(cudaMemcpy(h_sd2, d_sd2, n, cudaMemcpyDeviceToHost) );
- // 演算結果の確認
- for(i=0; i<nSize; i++){
- printf(" %f", h_sd2[i]);
- }
- printf("\n");
- // クリーンアップ
- free(h_sd1);
- free(h_sd2);
- CUDA_SAFE_CALL(cudaFree(d_sd1));
- CUDA_SAFE_CALL(cudaFree(d_sd2));
- }
- int
- main(int argc, char** argv)
- {
- runTest(argc, argv);
- CUT_EXIT(argc, argv);
- }
続いてGPUの処理を書いたgpukernel.cu
CU:
- // CPUから呼び出す関数
- // 配列に定数を掛けているだけの簡単なものです
- __global__ void
- test(float* fOut, float* fIn, int nSize, float f)
- {
- int i;
- for(i=0; i<nSize; i++){
- fOut[i] = fIn[i] * f;
- }
- }
ああ、これは楽チンだ。めんどくさい処理が全部消えてくれました。
コンパイルもnvccにcuda.cuを喰わせれば終了。例によってlibcutil.aとlibGLUT.aのリンクが必要だけど。
gridとthreadsという例のアレがまだ把握できていないのは問題として、とりあえずこれならサクサク組めそうです。
さて、もうちょっと色々といじってみるか。