OpenCL++!

1270803596|%Y-%m-%d|agohover

嗯,這就是我最近在做的事。

OpenCL 是一套在異質平台 (heterogeneous platforms) 上編寫平行處理程式的一套 API 標準。如果你平常就對什麼 CUDA 啦或是 GPGPU 之類的東西耳熟能詳,那應該也會聽過這個字眼,如果沒聽過的話就去參考一下 wiki 吧。

OpenCL 和 CUDA 的架構是很類似的,不過在程式編譯的流程上卻走了不同的方向。CUDA 本身是一個新的語言(儘管它非常像 C++),程式設計師使用 CUDA 的語法寫好程式後,要使用 NVIDIA 提供的 nvcc compiler 編譯自己的程式。nvcc 會把 CUDA 程式中給 GPU 執行的部份(也就是宣告為 __globale____device__ 的函式)取出來編譯成 GPU 所使用的 binary,然後把程式的其它部份轉換成合法的 C/C++ 後,交給其它的 compiler(VC 或是 GCC)來編譯,最後把這兩塊合起來。

OpenCL 的方式則是依照 OpenGL 的 shader,它定義了一組 C API 讓程式設計師可以編譯 device program。因此程式流程大致上是:

  1. 初始化 OpenCL context。
  2. 讀入 device program 的原始碼1並使用 OpenCL 提供的 API 進行編譯。
  3. 配置資料存取用的 device buffer。
  4. 設定 kernel function 的參數及 thread 總數並呼叫之。

以上的流程都是使用 C/C++ 呼叫 OpenCL API 來完成的,因此 OpenCL 的 host program 就只是一支單純的 C/C++ 程式。然而上述的前兩項步驟,在 CUDA 中都被 nvcc 自動幫你解決了。因此對懶惰的人來說,CUDA 是踏入 GPGPU 的一條捷徑。不過就我個人來說,是比較喜歡 OpenCL 的,因為:

  1. OpenCL 是標準。儘管這是一項很新的標準,但至少 GPU 兩大龍頭 NVIDIA 和 AMD 都會支援它。相較之下,CUDA 就只有 NVIDIA 的支援了。
  2. OpenCL 針對的不止是 GPGPU,而是提供異質平台的平行處理架構。因此除了 GPU 外,物理加速卡或是 cell processor 也是 OpenCL 可以使用的運算資源。
  3. 我比較喜歡 OpenCL 的 C extension。CUDA 的 C extension 雖然提供向量型別,但向量竟然不能加減乘除,難用爆了。而 OpenCL 的 device program 寫起來和 GLSL 很像,有好用的向量型別,當然 OpenCL 是設計成 general purpose,你不需要寫過 GLSL 也能很快就上手。

與 CUDA 比起來,OpenCL API 為了給使用者更大的彈性,它會顯得比較低階。以一個最簡單的例子來看:

// hello.cl
kernel void hello(global int *output)
{
    int i = get_global_id(0);
    output[i] = i*i;
}
// hello.cpp
#include <iostream>
#include <fstream>
#include <CL/cl.h>
 
using namespace std;
 
int main()
{
    // get the first OpenCL platform
    cl_platform_id platform;
    clGetPlatformIDs(1, &platform, NULL);
 
    // get the first GPU device on the first platform
    cl_device_id device;
    clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, 1, &device, NULL);
 
    // create OpenCL context
    cl_context_properties properties[] = {
        CL_CONTEXT_PLATFORM,
        reinterpret_cast<cl_context_properties>(platform),
        0
    };
    cl_context context;
    context = clCreateContext(properties, 1, &device, NULL, NULL, NULL);
 
    // read device program source
    ifstream fin("hello.cl", ios::binary);
    fin.seekg(0, ios::end);
    size_t len = fin.tellg();
    char* src = new char[len];
    fin.seekg(0, ios::beg);
    fin.read(&src[0], len);
 
    cl_program program;
    program = clCreateProgramWithSource(context, 1, const_cast<const char**>(&src), &len, NULL);
 
    // build the program
    clBuildProgram(program, 1, &device, NULL, NULL, NULL);
 
    // create a kernel object
    cl_kernel kernel = clCreateKernel(program, "hello", NULL);
 
    // create a buffer object
    cl_mem buf = clCreateBuffer(context, CL_MEM_READ_WRITE, sizeof(cl_int)*16, NULL, NULL);
 
    // set kernel arguments
    clSetKernelArg(kernel, 0, sizeof(cl_mem), &buf);
 
    // create the command queue object
    cl_command_queue queue = clCreateCommandQueue(context, device, 0, NULL);
 
    // execute the kernel
    size_t global_size = 16;
    size_t local_size = 4;
    cl_int err = clEnqueueNDRangeKernel(queue, kernel, 1, NULL, &global_size, &local_size, 0, NULL, NULL);
 
    // check the answer
    cl_int output[16];
    clEnqueueReadBuffer(queue, buf, CL_TRUE, 0, sizeof(cl_int)*16, output, 0, NULL, NULL);
 
    for(size_t i = 0; i < 16; ++i)
        cout << output[i] << ' ';
    cout << endl;
 
    // release resources
    delete [] src;
    clReleaseCommandQueue(queue);
    clReleaseMemObject(buf);
    clReleaseKernel(kernel);
    clReleaseProgram(program);
    clReleaseContext(context);
 
    return 0;
}

未免也太麻煩了,而且這還不包含錯誤處理!因此我嘗試用 C++ 來封裝 OpenCL API,同樣功能的 code 使用 C++ 的版本大概像這樣:

#include <iostream>
#include <exception>
 
#include "clpp/clpp.hpp"
 
using namespace std;
using namespace clpp;
 
int main()
{
    try{
        Context context;
 
        Kernel k = context.readProgramSourceFile("test.cl").kernel("hello");
 
        Buffer<cl_int> output = context.createBuffer<cl_int>(16);
        k.setArgs(output);
 
        CommandQueue q = context.queue();
        q.exec(k, 16, 4);
 
        // check the answer
        cl_int data[16];
        q.read(output, data);
 
        for(int i = 0; i < 16; ++i)
            cout << i << ' ';
        cout << endl;
    }catch(exception& err){
        cerr << err.what() << endl;
        return 1;
    }
    return 0;
}

當然,發生錯誤時會使用 exception 來進行妥善的回報。目前這個 clpp library 正在開發中,如果你有興趣的話也很歡迎來和我討論。


Comments

Add a New Comment
or Sign in as Wikidot user
(will not be published)
- +
c++
Unless otherwise stated, the content of this page is licensed under Creative Commons Attribution-ShareAlike 3.0 License