OpenCLで画像処理

今回はOpenCLを用いて画像処理を行ってみたいと思います。

下図のように、赤と青を入れ替える処理を実装していきます。

f:id:any-programming:20170319005947j:plain f:id:any-programming:20170319010003j:plain


アプリケーションコード

clCreateImageでもいいですが、今回はclCreateBufferで実装してみました。

BMPの読み書きは下記記事のものを利用しています。
Bitmap読み書き - 何でもプログラミング

#include <cl/cl.h>
#define ASSERT_CL(expr) if (expr != CL_SUCCESS) { throw std::exception(#expr); }

int main()
{
    // 画像読み込み
    int width, height;
    std::vector<byte> srcPixels;
    LoadBitmap24("Parrots.bmp", &width, &height, &srcPixels);

    // device, context作成
    cl_platform_id platform;
    ASSERT_CL(clGetPlatformIDs(1, &platform, nullptr));

    cl_device_id device;
    ASSERT_CL(clGetDeviceIDs(platform, CL_DEVICE_TYPE_ALL, 1, &device, nullptr));

    cl_context context = clCreateContext(nullptr, 1, &device, nullptr, nullptr, nullptr);
    assert(context);

    // kernel作成
    const char* source =
        "__kernel void main(__global const uchar* src, __global uchar *dst) {     \n"
        "   int i = get_global_id(0);                                             \n"
        "   dst[3 * i]     = src[3 * i + 2];                                      \n"
        "   dst[3 * i + 1] = src[3 * i + 1];                                      \n"
        "   dst[3 * i + 2] = src[3 * i];                                          \n"
        "}                                                                        \n";

    cl_program program = clCreateProgramWithSource(context, 1, &source, nullptr, nullptr);
    assert(program);

    ASSERT_CL(clBuildProgram(program, 1, &device, nullptr, nullptr, nullptr));
    char buildLog[1024];
    ASSERT_CL(clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_LOG, 1024, buildLog, nullptr));
    printf("build %s\n", buildLog);

    cl_kernel kernel = clCreateKernel(program, "main", nullptr);
    assert(kernel);

    // 引数設定
    int size = (int)srcPixels.size();
    cl_mem srcBuffer = clCreateBuffer(context, CL_MEM_READ_ONLY, size, nullptr, nullptr);
    assert(srcBuffer);
    cl_mem dstBuffer = clCreateBuffer(context, CL_MEM_WRITE_ONLY, size, nullptr, nullptr);
    assert(dstBuffer);

    ASSERT_CL(clSetKernelArg(kernel, 0, sizeof(srcBuffer), &srcBuffer));
    ASSERT_CL(clSetKernelArg(kernel, 1, sizeof(dstBuffer), &dstBuffer));

    // Queue作成
    cl_command_queue queue = clCreateCommandQueue(context, device, 0, nullptr);
    assert(queue);

    ASSERT_CL(clEnqueueWriteBuffer(queue, srcBuffer, CL_TRUE, 0, size, srcPixels.data(), 0, nullptr, nullptr));

    size_t global_work_size = size / 3;
    ASSERT_CL(clEnqueueNDRangeKernel(queue, kernel, 1, nullptr, &global_work_size, nullptr, 0, nullptr, nullptr));

    std::vector<byte> dstPixels(size);
    ASSERT_CL(clEnqueueReadBuffer(queue, dstBuffer, CL_TRUE, 0, size, dstPixels.data(), 0, nullptr, nullptr));

    // 実行
    ASSERT_CL(clFlush(queue));
    ASSERT_CL(clFinish(queue));

    // 画像保存
    SaveBitmap24("Parrots2.bmp", width, height, dstPixels.data());

    return 0;
}


SDKなしでの開発

OpenCLSDKは、IntellやNVIDIAなどから提供されています。

ちょっとしたものであれば、ヘッダファイルだけ取得して動的にdllをロードすることで開発できます。

ヘッダは下記よりダウンロードできます。(最低限、cl.h と cl_platform.h で大丈夫です。)
Khronos OpenCL Registry - The Khronos Group Inc

dllのロードは、下記コードのように行っています。

ヘッダに関数郡がすでに定義されているため、一段namespaceで包んで関数をロードしています。

#include <cl/cl.h>
namespace App {
    HMODULE hOpenCL = LoadLibraryA("opencl.dll");

#define GET_PROC(name) decltype(::name)* name = (decltype(::name)*)GetProcAddress(hOpenCL, #name)
    GET_PROC(clGetPlatformIDs);
    GET_PROC(clGetDeviceIDs);
    GET_PROC(clCreateContext);
    ...

    int main()
    {
        ....
    }
}

int main() { return App::main(); }