

### 実現したいこと

OpenCLで+1するプログラムを作成してみました。  
<https://peta.okechan.net/blog/archives/2538>  
を参考にしています。

[code]

    int index = get_global_id(0); 
    int index2 = get_global_id(1);
    
[/code]

を用いて `data[index*M+index2] += 1.0;` のようにアクセスしたいと思っています。

### 発生している問題

[code]

    int index = get_global_id(0); 
    int index2 = get_global_id(1); 
    
[/code]

を行うと常に`index`と`index2`に同じ値が格納されてしまいます。  
ほとんどの部分が+1されず終わっています。

### 該当のソースコード

[code]

    //
    //  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;
    }
    
[/code]

### 試したこと

[code]

    EC(clEnqueueNDRangeKernel(q, kernel, 2, offset, global, local, 0, nullptr, nullptr), "clEnqueueNDRangeKernel");
    
[/code]

のところの`global`, `local`などの数値を変えるなどして動かしてみました。

### 実行環境

MacBook Pro 13インチ、OSはel capitanです。

Terminal上で

[code]

    /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 
    
[/code]

で実行しました。


CC BY-SA 3.0
2017-01-29T04:20:10.880
32225
2017-01-29T17:33:28.313
2017-01-29T17:33:28.313
4548
20437
question
c++ c
OpenCLでの二次元配列の処理の仕方について


OpenCLの実行環境がないので実行の確認はできていないんですが，カーネルの実行の部分

[code]

    size_t global[2],local[2],offset[2];
    offset[0] = 0;
    offset[0] = 0; //<- 本当は offset[1] = 0 では？
    
[/code]

となっています．これでは`offset[1]`が未初期化なので動作が不定になるものと考えられます．

こういった初期化漏れを防ぐために，初期化はなるべく宣言時に行うべきです．  
今回の場合であれば，

[code]

    size_t offset[2] = {0, 0};
    
[/code]

などとすればいいかと思います．


CC BY-SA 3.0
2017-01-29T10:32:41.663
32227
2017-01-29T14:38:17.290
2017-01-29T14:38:17.290
4548
4548
32225
answer