###前提・実現したいこと
OpenCLで+1するプログラムを作成してみました。
https://peta.okechan.net/blog/archives/2538
を参考にしています。
int index = get_global_id(0);
int index2 = get_global_id(1);
を用いて
data[index*M+index2] += 1.0;
のようにアクセスしたいと思っています。
###発生している問題・エラーメッセージ
int index = get_global_id(0);
int index2 = get_global_id(1);
を行うと常にindexとindex2に同じ値が格納されてしまいます。
ほとんどの部分が+1されず終わっています。
###該当のソースコード // // main3.cpp // #include <iostream> #include <vector> #include <OpenCL/opencl.h> #include <numeric> #define PLATFORM_MAX 4 #define DEVICE_MAX 4 void EC(cl_int result, const char *title) { if (result != CL_SUCCESS) { std::cout << "Error: " << title << "(" << result << ")\n"; } } cl_int err = CL_SUCCESS; void EC2(const char *title) { if (err != CL_SUCCESS) { std::cout << "Error: " << title << "(" << err << ")\n"; } err = CL_SUCCESS; } int main(int argc, const char * argv[]) { // プラットフォーム一覧を取得 cl_platform_id platforms[PLATFORM_MAX]; cl_uint platformCount; EC(clGetPlatformIDs(PLATFORM_MAX, platforms, &platformCount), "clGetPlatformIDs"); if (platformCount == 0) { std::cerr << "No platform.\n"; return EXIT_FAILURE; } // 見つかったプラットフォームの情報を印字 for (int i = 0; i < platformCount; i++) { char vendor[100] = {0}; char version[100] = {0}; EC(clGetPlatformInfo(platforms[i], CL_PLATFORM_VENDOR, sizeof(vendor), vendor, nullptr), "clGetPlatformInfo"); EC(clGetPlatformInfo(platforms[i], CL_PLATFORM_VERSION, sizeof(version), version, nullptr), "clGetPlatformInfo"); std::cout << "Platform id: " << platforms[i] << ", Vendor: " << vendor << ", Version: " << version << "\n"; } // デバイス一覧を取得 cl_device_id devices[DEVICE_MAX]; cl_uint deviceCount; EC(clGetDeviceIDs(platforms[0], CL_DEVICE_TYPE_GPU, DEVICE_MAX, devices, &deviceCount), "clGetDeviceIDs"); if (deviceCount == 0) { std::cerr << "No device.\n"; return EXIT_FAILURE; } // 見つかったデバイスの情報を印字 std::cout << deviceCount << " device(s) found.\n"; for (int i = 0; i < deviceCount; i++) { char name[100] = {0}; size_t len; EC(clGetDeviceInfo(devices[i], CL_DEVICE_NAME, sizeof(name), name, &len), "clGetDeviceInfo"); std::cout << "Device id: " << i << ", Name: " << name << "\n"; } // コンテキストの作成 cl_context ctx = clCreateContext(nullptr, 1, devices, nullptr, nullptr, &err); EC2("clCreateContext"); // コンパイル済みclプログラムの読み込み const char* bitcode_path = "kernel2.cl.gpu_32.bc"; size_t len = strlen(bitcode_path); cl_program program = clCreateProgramWithBinary(ctx, 1, devices, &len, (const unsigned char**)&bitcode_path, nullptr, &err); EC2("clCreateProgramWithBinary"); // プログラムのビルド EC(clBuildProgram(program, 1, devices, nullptr, nullptr, nullptr), "clBuildProgram"); // カーネルの作成 cl_kernel kernel = clCreateKernel(program, "addone", &err); EC2("clCreateKernel"); // データを用意 int n = 10; std::vector<float> data(n*n,0.0f); // デバイスメモリを確保しつつデータをコピー cl_mem device_mem = clCreateBuffer(ctx, CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR, sizeof(float) * n*n, data.data(), &err); EC2("clCreateBuffer"); // カーネルの引数をセット EC(clSetKernelArg(kernel, 0, sizeof(cl_mem), &device_mem), "clSetKernelArg"); EC(clSetKernelArg(kernel, 1, sizeof(int), &n), "clSetKernelArg"); // コマンドキューの作成 cl_command_queue q = clCreateCommandQueue(ctx, devices[0], 0, &err); EC2("clCreateCommandQueue"); // カーネルの実行 size_t global[2],local[2],offset[2]; offset[0] = 0; offset[0] = 0; global[0] = n; global[1] = n; local[0] = 1; local[1] = 1; EC(clEnqueueNDRangeKernel(q, kernel, 2, offset, global, NULL, 0, nullptr, nullptr), "clEnqueueNDRangeKernel"); // 結果を読み込み EC(clEnqueueReadBuffer(q, device_mem, CL_TRUE, 0, sizeof(float) * n*n, data.data(), 0, nullptr, nullptr), "clEnqueueReadBuffer"); // 結果の印字 for (int i = 0; i < n*n; i++) { std::cout << data[i] << ", "; } std::cout << "\n"; float total = std::accumulate(data.begin(),data.end(),0.0); std::cout << total << std::endl; // コマンドキューの解放 EC(clReleaseCommandQueue(q), "clReleaseCommandQueue"); // デバイスメモリを解放 EC(clReleaseMemObject(device_mem), "clReleaseMemObject"); // カーネルの解放 EC(clReleaseKernel(kernel), "clReleaseKernel"); // プログラムの解放 EC(clReleaseProgram(program), "clReleaseProgram"); // コンテキストの解放 EC(clReleaseContext(ctx), "clReleaseContext"); std::cout << "Done.\n"; return EXIT_SUCCESS; } //ここからカーネル部分 //kernel2.cl __kernel void addone(__global float* data,const int n) { int index = get_global_id(0); int index2 = get_global_id(1); int dim = get_work_dim(); printf("get_work_dim = %d\n",dim); printf("index = %d , index2 = %d \n",index,index2); data[index*n+index2] += 1.0f; } ```C++
###試したこと
EC(clEnqueueNDRangeKernel(q, kernel, 2, offset, global, local, 0, nullptr, nullptr), "clEnqueueNDRangeKernel");のところのglobal,localなどの数値を変えるなどして動かしてみました。
###補足情報(言語/FW/ツール等のバージョンなど)
MacBook Pro 13インチ、OSはel capitanです。
Terminal上で
/System/Library/Frameworks/OpenCL.framework/Libraries/openclc -c -o kernel2.cl.gpu_32.bc -arch gpu_32 -emit-llvm kernel2.cl
g++ -O3 -std=c++11 -framework opencl main3.cpp -o test
./test
で実行しました。
バッドをするには、ログインかつ
こちらの条件を満たす必要があります。
2017/01/11 01:32
2017/01/11 01:35
2017/01/11 01:46
2017/01/11 02:08
2017/01/11 02:29 編集
2017/01/11 03:22
2017/01/11 03:28
2017/01/11 03:43