IntelのOpenCLドライバもあるらしいので、インストールした流れで、AMD RadeonとCore i7両方を使って計算をしてみることにした。
そのまえに、ちゃんとIntelドライバを認識しているか、AMD Accelerated Parallel Processing (APP) SDKに付属している"clinfo.exe"で確認した。
C:\Program Files (x86)\AMD APP\bin\x86_64>clinfo.exe Number of platforms: 2 Platform Profile: FULL_PROFILE Platform Version: OpenCL 1.1 AMD-APP (831.4) Platform Name: AMD Accelerated Parallel Proces sing Platform Vendor: Advanced Micro Devices, Inc. Platform Extensions: cl_khr_icd cl_amd_event_callbac k cl_amd_offline_devices cl_khr_d3d10_sharing Platform Profile: FULL_PROFILE Platform Version: OpenCL 1.1 Platform Name: Intel(R) OpenCL Platform Vendor: Intel(R) Corporation Platform Extensions: cl_khr_fp64 cl_khr_global_int32 _base_atomics cl_khr_global_int32_extended_atomics cl_khr_local_int32_base_atomi cs cl_khr_local_int32_extended_atomics cl_khr_byte_addressable_store cl_intel_pr intf cl_ext_device_fission cl_intel_immediate_execution cl_khr_gl_sharing cl_khr _icd
こんな感じの出力が出て、確かにIntel提供のプラットフォームが増えている。
「OpenCLで実際に計算してみる」のコードを単純にAMD用とIntel用に二重化して動かしてみた。
#include <cstdio> #include <cmath> #include <boost/chrono.hpp> #include <boost/chrono/duration.hpp> using namespace boost; #ifdef __APPLE__ #include <OpenCL/opencl.h> #else #include <CL/cl.h> #endif //__APPLE__ const int nElements = 9000000; const int maxDevices = 10; float input1[nElements]; float input2[nElements]; float output[nElements]; void addVector(float* input1, float* input2, float* output){ for(int i = 0; i < nElements; i++){ output[i] = sin(input1[i]) * sin(input2[i]); output[i] = cos(output[i]); output[i] = pow(output[i], output[i]); } } int main(int argc, char* argv[]) { cl_int status; cl_platform_id platforms[10]; cl_uint num_platforms; status = clGetPlatformIDs(sizeof(platforms) / sizeof(platforms[0]), platforms, &num_platforms); std::cout << "Number Of Platforms : " << num_platforms << std::endl; cl_context_properties amdProp[] = {CL_CONTEXT_PLATFORM, (cl_context_properties)platforms[0], 0}; cl_context_properties intelProp[] = {CL_CONTEXT_PLATFORM, (cl_context_properties)platforms[1], 0}; cl_context amdContext = clCreateContextFromType(amdProp, CL_DEVICE_TYPE_GPU, NULL, NULL, &status); cl_context intelContext = clCreateContextFromType(intelProp, CL_DEVICE_TYPE_CPU, NULL, NULL, &status); cl_device_id amdDev[maxDevices]; size_t size_return; status = clGetContextInfo(amdContext, CL_CONTEXT_DEVICES, sizeof(amdDev), amdDev, &size_return); cl_device_id intelDev[maxDevices]; status = clGetContextInfo(intelContext, CL_CONTEXT_DEVICES, sizeof(intelDev), intelDev, &size_return); cl_command_queue amdQueue = clCreateCommandQueue(amdContext, amdDev[0], 0, &status); cl_command_queue intelQueue = clCreateCommandQueue(intelContext, intelDev[0], 0, &status); const char *sources[] = { "__kernel void\n\ addVector(__global const float *input1,\n\ __global const float *input2,\n\ __global float *output)\n\ {\n\ int index = get_global_id(0);\n\ output[index] = sin(input1[index]) * sin(input2[index]);\n\ output[index] = cos(output[index]);\n\ output[index] = pow(output[index], output[index]);\n\ }\n"}; cl_program amdProgram = clCreateProgramWithSource(amdContext, 1, (const char**)&sources, NULL, &status); cl_program intelProgram = clCreateProgramWithSource(intelContext, 1, (const char**)&sources, NULL, &status); status = clBuildProgram(amdProgram, 1, amdDev, NULL, NULL, NULL); status = clBuildProgram(intelProgram, 1, intelDev, NULL, NULL, NULL); clUnloadCompiler(); cl_kernel amdKernel = clCreateKernel(amdProgram, "addVector", &status); cl_kernel intelKernel = clCreateKernel(intelProgram, "addVector", &status); for (int i = 0; i < nElements; i++) { input1[i] = (float)i * 10.0f; input2[i] = (float)i / 20.0f; output[i] = 0.0f; } cl_mem amdMemInput1 = clCreateBuffer(amdContext, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, sizeof(cl_float) * nElements, input1, &status); cl_mem amdMemInput2 = clCreateBuffer(amdContext, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, sizeof(cl_float) * nElements, input2, &status); cl_mem amdMemOutput = clCreateBuffer(amdContext, CL_MEM_WRITE_ONLY, sizeof(cl_float) * nElements, NULL, &status); cl_mem intelMemInput1 = clCreateBuffer(intelContext, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, sizeof(cl_float) * nElements, input1, &status); cl_mem intelMemInput2 = clCreateBuffer(intelContext, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, sizeof(cl_float) * nElements, input2, &status); cl_mem intelMemOutput = clCreateBuffer(intelContext, CL_MEM_WRITE_ONLY, sizeof(cl_float) * nElements, NULL, &status); status = clSetKernelArg(amdKernel, 0, sizeof(cl_mem), (void *)&amdMemInput1); status = clSetKernelArg(amdKernel, 1, sizeof(cl_mem), (void *)&amdMemInput2); status = clSetKernelArg(amdKernel, 2, sizeof(cl_mem), (void *)&amdMemOutput); status = clSetKernelArg(intelKernel, 0, sizeof(cl_mem), (void *)&intelMemInput1); status = clSetKernelArg(intelKernel, 1, sizeof(cl_mem), (void *)&intelMemInput2); status = clSetKernelArg(intelKernel, 2, sizeof(cl_mem), (void *)&intelMemOutput); boost::chrono::steady_clock::time_point start = boost::chrono::steady_clock::now(); size_t globalSize[] = {nElements}; status = clEnqueueNDRangeKernel(amdQueue, amdKernel, 1, NULL, globalSize, 0, 0, NULL, NULL); status = clEnqueueReadBuffer(amdQueue, amdMemOutput, CL_TRUE, 0, sizeof(cl_float) * nElements, output, 0, NULL, NULL); boost::chrono::duration sec = boost::chrono::steady_clock::now() - start; std::cout << "input1, input2, output" << std::endl; for (int i = 0; i < 100; i++) { std::cout << input1[i] << ", " << input2[i] << ", " << output[i] << std::endl; } std::cout << "OpenCL addVector(AMD) " << sec.count() << " seconds\n"; boost::chrono::steady_clock::time_point start_i = boost::chrono::steady_clock::now(); status = clEnqueueNDRangeKernel(intelQueue, intelKernel, 1, NULL, globalSize, 0, 0, NULL, NULL); status = clEnqueueReadBuffer(intelQueue, intelMemOutput, CL_TRUE, 0, sizeof(cl_float) * nElements, output, 0, NULL, NULL); boost::chrono::duration sec_i = boost::chrono::steady_clock::now() - start_i; std::cout << "input1, input2, output" << std::endl; for (int i = 0; i < 100; i++) { std::cout << input1[i] << ", " << input2[i] << ", " << output[i] << std::endl; } std::cout << "OpenCL addVector(intel) " << sec_i.count() << " seconds\n"; clReleaseMemObject(amdMemOutput); clReleaseMemObject(amdMemInput2); clReleaseMemObject(amdMemInput1); clReleaseKernel(amdKernel); clReleaseProgram(amdProgram); clReleaseCommandQueue(amdQueue); clReleaseContext(amdContext); clReleaseMemObject(intelMemOutput); clReleaseMemObject(intelMemInput2); clReleaseMemObject(intelMemInput1); clReleaseKernel(intelKernel); clReleaseProgram(intelProgram); clReleaseCommandQueue(intelQueue); clReleaseContext(intelContext); boost::chrono::steady_clock::time_point start2 = boost::chrono::steady_clock::now(); addVector(input1, input2, output); boost::chrono::duration sec2 = boost::chrono::steady_clock::now() - start2; std::cout << "Normal addVector " << sec2.count() << " seconds\n"; return 0; }
amd*はRadeon用、intel*はCore i7用として、それぞれ実行時間を計測した。
実行時間部分の結果は以下のようになる。
OpenCL addVector(AMD) 0.129915 seconds OpenCL addVector(intel) 0.114104 seconds Normal addVector 1.12876 seconds
何回か実行したが、若干Intelプラットフォームの方が実行時間は短かった。
マルチスレッド化して、スレッドごとにデバイスを割り当てるなんてこともできそうな気がする。このあたりの結果を「OpenCLをクラスでラップしてみた」のクラスにでも反映していこうかなと思う。
コメント