前回はSDKなしの状態で、OpenCLの関数を実行して、OpenCL実行の際のデバイスの情報を取得する部分まで行いました。今回はOpenCLのコンテキストを生成して配列の加算くらいをやってみたいと思います。
前回の補足
前回はAMDグラフィックボードが搭載された環境での実行結果を掲載しました。あれからNVIDIAボードを搭載した環境でも実行してみたので、結果をここに貼っておきたいと思います。
PlatformName: NVIDIA CUDA DeviceName: GeForce GTX 650 Ti > DeviceExtensions: cl_khr_byte_addressable_store cl_khr_icd cl_khr_gl_sharing cl_nv_d3d9_sharing cl_nv_d3d10_sharing cl_khr_d3d10_sharing cl_nv_d3d11_sharing cl_nv_compiler_options cl_nv_device_attribute_query cl_nv_pragma_unroll cl_khr _global_int32_base_atomics cl_khr_global_int32_extended_atomics cl_khr_local_int 32_base_atomics cl_khr_local_int32_extended_atomics cl_khr_fp64 PlatformName: Intel(R) OpenCL DeviceName: Intel(R) Core(TM) i7-3770T CPU @ 2.50GHz > DeviceExtensions: cl_khr_fp64 cl_khr_icd cl_khr_global_int32_base_atomics cl _khr_global_int32_extended_atomics cl_khr_local_int32_base_atomics cl_khr_local_ int32_extended_atomics cl_khr_byte_addressable_store cl_intel_printf cl_ext_devi ce_fission cl_khr_gl_sharing cl_intel_dx9_media_sharing
意外にもCPUはこちらでも使えるようでした。最悪ケースとしてもCPUでOpenCLが利用可能であるのかもしれません。また、この環境ではIntel HD Graphics 4000も有効化してあるのですが、この項目は上記の部分で出てきませんでした。
ドライバも最新の状態で、ディスプレイ未接続でQSVが使えるのを確認しています。それなのに利用可能デバイスとして出てこなかったのは意外でした。
初期化処理
ここでは既にOpenCLに関して知識があるものとして、個別のAPI説明は行いません。
関数ポインタを取り出して OpenCLカーネルの準備まで行うコードを以下に示します。やっていることは通常のOpenCLを使うまでの手順と変わりません。
// 関数ポインタ. 事前に型は定義しておく PFNCLCREATECONTEXT _g_clCreateContext = NULL; PFNCLRELEASECONTEXT _g_clReleaseContext = NULL; PFNCLCREATECOMMANDQUEUE _g_clCreateCommandQueue = NULL; PFNCLRELEASECOMMANDQUEUE _g_clReleaseCommandQueue = NULL; PFNCLCREATEBUFFER _g_clCreateBuffer = NULL; PFNCLRELEASEMEMOBJECT _g_clReleaseMemObject = NULL; PFNCLCREATEPROGRAMWITHSOURCE _g_clCreateProgramWithSource = NULL; PFNCLRELEASEPROGRAM _g_clReleaseProgram = NULL; PFNCLBUILDPROGRAM _g_clBuildProgram = NULL; PFNCLCREATEKERNEL _g_clCreateKernel = NULL; PFNCLRELEASEKERNEL _g_clReleaseKernel = NULL; PFNCLSETKERNELARG _g_clSetKernelArg = NULL; PFNCLENQUEUEREADBUFFER _g_clEnqueueReadBuffer = NULL; PFNCLENQUEUEWRITEBUFFER _g_clEnqueueWriteBuffer = NULL; PFNCLENQUEUENDRANGEKERNEL _g_clEnqueueNDRangeKernel = NULL; PFNCLFINISH _g_clFinish = NULL; // 関数ポインタの取り出し部 _g_clCreateContext = (PFNCLCREATECONTEXT)GetProcAddress( h, "clCreateContext" ); _g_clReleaseContext = (PFNCLRELEASECONTEXT)GetProcAddress( h, "clReleaseContext" ); _g_clCreateCommandQueue = (PFNCLCREATECOMMANDQUEUE)GetProcAddress( h, "clCreateCommandQueue" ); _g_clReleaseCommandQueue = (PFNCLRELEASECOMMANDQUEUE)GetProcAddress( h, "clReleaseCommandQueue" ); _g_clCreateBuffer = (PFNCLCREATEBUFFER)GetProcAddress( h, "clCreateBuffer" ); _g_clReleaseMemObject = (PFNCLRELEASEMEMOBJECT)GetProcAddress( h, "clReleaseMemObject" ); _g_clCreateProgramWithSource = (PFNCLCREATEPROGRAMWITHSOURCE) GetProcAddress( h, "clCreateProgramWithSource" ); _g_clBuildProgram = (PFNCLBUILDPROGRAM) GetProcAddress( h, "clBuildProgram" ); _g_clReleaseProgram = (PFNCLRELEASEPROGRAM) GetProcAddress( h, "clReleaseProgram" ); _g_clCreateKernel = (PFNCLCREATEKERNEL) GetProcAddress( h, "clCreateKernel" ); _g_clReleaseKernel = (PFNCLRELEASEKERNEL)GetProcAddress( h, "clReleaseKernel" ); _g_clSetKernelArg = (PFNCLSETKERNELARG) GetProcAddress( h, "clSetKernelArg" ); _g_clEnqueueNDRangeKernel = (PFNCLENQUEUENDRANGEKERNEL) GetProcAddress( h, "clEnqueueNDRangeKernel" ); _g_clFinish = (PFNCLFINISH)GetProcAddress( h, "clFinish" ); _g_clEnqueueReadBuffer = (PFNCLENQUEUEREADBUFFER) GetProcAddress( h, "clEnqueueReadBuffer" ); _g_clEnqueueWriteBuffer = (PFNCLENQUEUEWRITEBUFFER) GetProcAddress( h, "clEnqueueWriteBuffer" );
// カーネルの準備までを行う
cl_platform_id platformId = NULL;
cl_device_id deviceId = NULL;
_g_clGetPlatformIDs( 1, &platformId, NULL );
retCl = _g_clGetDeviceIDs( platformId, CL_DEVICE_TYPE_GPU, 1, &deviceId, NULL );
if( retCl != CL_SUCCESS ) {
return -1;
}
cl_context contextCL = _g_clCreateContext( 0, 1, &deviceId, NULL, NULL, &retCl );
if( retCl != CL_SUCCESS ) {
return -1;
}
cl_command_queue commands = _g_clCreateCommandQueue( contextCL, deviceId, NULL, &retCl );
if( retCl != CL_SUCCESS ) {
return -1;
}
size_t programLength = strlen( kernelSrc );
cl_program program = _g_clCreateProgramWithSource( contextCL, 1, (const char**)&kernelSrc, &programLength, &retCl );
if( retCl != CL_SUCCESS ) {
return -1;
}
retCl = _g_clBuildProgram( program, 1, &deviceId, NULL, NULL, NULL );
if( retCl != CL_SUCCESS ) {
return -1;
}
cl_kernel kernel = _g_clCreateKernel( program, "vector_add", &retCl );
if( retCl != CL_SUCCESS ) {
return -1;
}
配列加算用の処理
CLのバッファを合計３つ用意してキューに追加して、カーネルを実行します。
カーネルは下記に示すコードです。今回はコード中に直接文字列として埋め込んでいます。
const char* kernelSrc =
"__kernel void vector_add(__global const int* inputA, __global const int* inputB, __global int* outputC ){\n"
" int idx = get_global_id(0);\n"
" outputC[idx] = inputA[idx] + inputB[idx];\n"
"}\0";
const int BUFFER_SIZE = 1024 * 1024;
const int BUFFER_BYTE_SIZE = sizeof(int) * BUFFER_SIZE;
cl_mem bufA = _g_clCreateBuffer( contextCL, CL_MEM_READ_ONLY, BUFFER_BYTE_SIZE, NULL, &retCl );
cl_mem bufB = _g_clCreateBuffer( contextCL, CL_MEM_READ_ONLY, BUFFER_BYTE_SIZE, NULL, &retCl );
cl_mem bufC = _g_clCreateBuffer( contextCL, CL_MEM_WRITE_ONLY, BUFFER_BYTE_SIZE, NULL, &retCl );
int* hostArrayA = (int*)malloc( BUFFER_BYTE_SIZE );
int* hostArrayB = (int*)malloc( BUFFER_BYTE_SIZE );
int* hostArrayC = (int*)malloc( BUFFER_BYTE_SIZE );
memset( hostArrayC, 0, BUFFER_BYTE_SIZE);
for( int i = 0; i < BUFFER_SIZE; ++i ) {
hostArrayA[i] = i;
hostArrayB[i] = BUFFER_SIZE - i;
}
_g_clEnqueueWriteBuffer( commands, bufA, CL_FALSE, 0, BUFFER_BYTE_SIZE, hostArrayA, 0, NULL, NULL );
_g_clEnqueueWriteBuffer( commands, bufB, CL_FALSE, 0, BUFFER_BYTE_SIZE, hostArrayB, 0, NULL, NULL );
_g_clSetKernelArg( kernel, 0, sizeof(cl_mem), &bufA );
_g_clSetKernelArg( kernel, 1, sizeof(cl_mem), &bufB );
_g_clSetKernelArg( kernel, 2, sizeof(cl_mem), &bufC );
size_t arraySize = BUFFER_SIZE;
_g_clEnqueueNDRangeKernel( commands, kernel, 1, NULL, &arraySize, NULL, 0, NULL, NULL );
_g_clEnqueueReadBuffer( commands, bufC, CL_TRUE, 0, BUFFER_BYTE_SIZE, hostArrayC, 0, NULL, NULL );
_g_clFinish( commands );
for( int i = 0; i < 10; ++i ) {
printf( " %d : %d\n", i, hostArrayC[i] );
}
後始末
使ったりソースは解放します。
_g_clReleaseKernel( kernel ); _g_clReleaseProgram( program ); _g_clReleaseMemObject( bufA ); _g_clReleaseMemObject( bufB ); _g_clReleaseMemObject( bufC ); _g_clReleaseCommandQueue( commands ); _g_clReleaseContext( contextCL ); free( hostArrayA ); free( hostArrayB ); free( hostArrayC );
まとめと注意事項
とりあえず今回のコードで配列の加算ができるところまでやってみました。ここまでの動作では確かにSDKのインストール無しで試してみることが出来ました。どうやらうまくいきそうな気配ではあります。ただし、関数ポインタの取り出しがやっぱり面倒には感じます。一度この部分のラッパーを書いてしまえば問題ないのでしょうが、そのくらいならSDKの導入というのもアリな気がしてきます。おそらくはSDKを導入しての開発の方が自然でしょうし、今回のDLLから取り出して初期化して実行するという方法も１つの手ではあろうと思いますが、イマイチ正式な方法とも言いがたいと思っています。
思い立ったのでやってみましたが、この方法をオススメするという意図でブログに書いたわけでもなく、単に”できたよ”の意味でやってみました。もしこの方法で何かを実装していく際には検証のほどを忘れずに。当方ではうまくいかなくても責任は負いかねます・・・。(一応 AMD, NVIDIAの搭載環境でチェックくらいはしましたが・・・)
