OpenCLのC++ Bindingsを使ってみる - その5

「OpenCLのC++ Bindingsを使ってみる - その1」、「OpenCLのC++ Bindingsを使ってみる - その2」、「OpenCLのC++ Bindingsを使ってみる - その3」、「OpenCLのC++ Bindingsを使ってみる - その4」に引き続き、今回はOpenCLのC++ Bindingsを使って、コマンドキューの生成とキューにカーネルを入れ実際に計算することにする。これまでと同様に前回までのコードに追記していく。

 #define __CL_ENABLE_EXCEPTIONS
 
 #if defined(__APPLE__) || defined(__MACOSX)
    #include 
 #else
    #include 
 #endif
 #include 
const int nElements = 9000000;
float input1[nElements];
float input2[nElements];
float output[nElements];
int main(int argc, char* argv[])
{
    cl_int error = CL_SUCCESS;

    try{
        std::vector platforms;
        cl::Platform::get(&platforms);
        if(platforms.size() == 0){
            std::cout << "Any Platforms is NOT FOUNT." << std::endl;
            return 1;
        }

        cl_context_properties properties[] =
            {CL_CONTEXT_PLATFORM, (cl_context_properties)(platforms[0])(), 0};
        cl::Context context(CL_DEVICE_TYPE_GPU, properties);
        std::vector devices = context.getInfo();

        char* addVector = {
        "__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::Sources source(1,std::make_pair(addVector, strlen(addVector)));
        cl::Program program = cl::Program(context, source);
        program.build(devices);
        cl::Kernel kernel(program, "addVector", &error);

        for(int i = 0; i < nElements; i++){
            input1[i] = (float)i * 10.0f;
            input2[i] = (float)i / 20.0f;
            output[i] = 0.0f;
        }

        cl::Buffer memInput1 = cl::Buffer(context,
            CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR,
            sizeof(cl_float) * nElements,
            input1,
            &error);
        cl::Buffer memInput2 = cl::Buffer(context,
            CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR,
            sizeof(cl_float) * nElements,
            input2,
            &error);
        cl::Buffer memOutput = cl::Buffer(context,
            CL_MEM_WRITE_ONLY,
            sizeof(cl_float) * nElements,
            NULL,
            &error);

        kernel.setArg(0, memInput1);
        kernel.setArg(1, memInput2);
        kernel.setArg(2, memOutput);
        
        cl::CommandQueue queue(context, devices.at(0), 0, &error);
        queue.enqueueNDRangeKernel(kernel, cl::NullRange,
            cl::NDRange(nElements), cl::NullRange, NULL, NULL);

        queue.enqueueReadBuffer(memOutput, CL_TRUE, 0,
            sizeof(cl_float) * nElements, output, NULL, NULL);

        for(int i = 0; i < 20; i++){
            std::cout << "input1[" << i << "], input2[" << i << "], output[" << i << "] : ";
            std::cout << input1[i] << ", " << input2[i] << ", " << output[i] << std::endl;
        }
    }catch(cl::Error err){
         std::cerr << "ERROR: " << err.what() << "(" << err.err() << ")" << std::endl;
    }

    return 0;
}

まず、cl::CommandQueueクラスのインスタンスを生成する。この時の引数はコンテキストとデバイス、キューのプロパティ、エラーコードとなる。キューのプロパティはOpenCL 1.1 Specification [PDF]のTable 5.1に記載されている。queue.enqueueNDRangeKernelでは、生成したコマンドキューにカーネルを入れている。この時の引数はカーネル、オフセット、グローバルデータの次元、ローカルデータの次元、イベントリスト、イベント出力用変数となる。ここで、カーネルが実行される。

次に結果を取得する。queue.enqueueReadBufferで出力用変数outputに結果を読み込む。最後に読み込んだ変数の一部を表示している。

結果このようになる

input1[0], input2[0], output[0] : 0, 0, 1
input1[1], input2[1], output[1] : 10, 0.05, 0.999631
input1[2], input2[2], output[2] : 20, 0.1, 0.995867
input1[3], input2[3], output[3] : 30, 0.15, 0.989237
input1[4], input2[4], output[4] : 40, 0.2, 0.989182
input1[5], input2[5], output[5] : 50, 0.25, 0.997898
input1[6], input2[6], output[6] : 60, 0.3, 0.995962
input1[7], input2[7], output[7] : 70, 0.35, 0.966201
input1[8], input2[8], output[8] : 80, 0.4, 0.93131
input1[9], input2[9], output[9] : 90, 0.45, 0.930719
input1[10], input2[10], output[10] : 100, 0.5, 0.971525
input1[11], input2[11], output[11] : 110, 0.55, 0.999733
input1[12], input2[12], output[12] : 120, 0.6, 0.949504
input1[13], input2[13], output[13] : 130, 0.65, 0.867867
input1[14], input2[14], output[14] : 140, 0.7, 0.841198
input1[15], input2[15], output[15] : 150, 0.75, 0.896425
input1[16], input2[16], output[16] : 160, 0.8, 0.987789
input1[17], input2[17], output[17] : 170, 0.85, 0.967397
input1[18], input2[18], output[18] : 180, 0.9, 0.842729
input1[19], input2[19], output[19] : 190, 0.95, 0.7733

これで、OpenCL C++ Bindingsを使用した一連の計算をすることができた。

このブログの開発用PCはこちら