LoginSignup
14
15

More than 5 years have passed since last update.

TinyCLというものを作りました

Last updated at Posted at 2014-08-21

はじめに

どうやらOpenCLはベンダーによって提供しているライブラリが微妙に違うらしい.NVIDIAでは,CL/cl.hppが存在しないため,C++でOpenCLを使うことができない.となると,CL/cl.hppを利用すると,もれなくNVIDIAでは動かなくなる可能性が高い,ということになる.

それはちょっとアカンやろ,と思いまして,どうせならモダンなC++で,かつ短くわかりやすく簡単に書けるようなOpenCLのラッパーを作っちゃえばいいんじゃないかな,って作ってみました.

NVIDIAドライバ + Visual Studioの導入方法はこちらです.
http://qiita.com/GRGSIBERIA/items/d8290c1a3ded59f4075a

使い方

使い方は非常にシンプルです.

カーネルのソースコードは下です.
ちなみにget_local_id関数は,スレッド番号を取得しています.

__kernel void TestMain(
    __global float* inputA, __global float* inputB)
{
    int gid = get_local_id(0);
    inputA[gid] = gid;
    inputB[gid] = gid * gid;
}

TinyCLを使っている様子です.

#include <iostream>
#include "TinyCL.hpp"

int main()
{
    // デバイスに渡すための配列を0で初期化
    const size_t N = 10;
    std::vector<float> inputA(N, 0), inputB(N, 0);

    // ソースコードのコンパイル
    tcl::CLController controller("test.cl", "TestMain");

    // デバイス側のメモリを確保
    tcl::CLReadWriteBuffer bufA(inputA), bufB(inputB);

    // N個のワーカーを用意し,引数を設定して実行,Result関数で結果を書き込む
    controller.Setting(N).Run(bufA, bufB).Result();

    // 中身が正しいかどうか確認する
    for (int i = 0; i < N; ++i)
        std::cout << i << "," << inputA[i] << "," << inputB[i] << std::endl;

    return 0;
}

細かいところはコメントで.NVIDIAしか実行の確認できないのであしからず.他の環境を持ってる方がいたら,動かない部分をDEFINEとかで逃してください.

2015/02/19追記

ソースコードをもっと短くしてみました.以前のものと比較するとわかるのですが,CLControllerクラスを新しく作り,細かい設定をハンドリングするようにしました.これで最小4行程度でOpenCLを使えるようになります.

以前はOpenCLの実装になるべく忠実な形で抽象化していましたが,久しぶりに使ってみると使い方を忘れてしまいました.これだとさすがにアカンということで,思い出せるレベルで簡単なAPIにしたかったので,CLControllerを追加しました.一応,色々やりたい人向けの機能も残してあります.

2015/02/21追記

引数を2つ以上指定すると動かないという,かなり深刻なバグがあったので修正しました.

2015/02/23追記

たったの3行で使えるようになりました.

解説

なるべく短く書けるように工夫しました.CLBufferはコンストラクタに配列を渡せるようにしておけばもう少し短くなるかもしれません.

OpenCLはC言語で書かれているので,そこかしこにポインタを使っています.
特に,変数の型がvoid*なのに,引数の都合でvoid**にキャストしないと困るところが多いです.
変換ミスって落ちることもあります.

また,OpenCLは関数が少ない代わりに,フラグやvoid*を引数に渡して管理するので非常に危険です.
VC++でも,インテリセンスに引数名が出なくて苦しみました.
普通にリファレンスが手元に置いて,一つ一つ引数を目視でチェックしないとヤバイです.
N番目からスタートして,0個のクラスタをN個で分けるとか,これ見つけるのに2日かかりました.

このあたりのミスを誘発するような設計では,とてもとてもコードを書いていられません.
なので,なるべくミスらないように,しかも綺麗に少ない行数で書けるよう,工夫することで解決しました.

基本的には,デバイスの取得やカーネルの生成などは,本当に基本的なことしかやらなければ,定形処理で済みます.
このような部分は積極的にカプセル化することで,OpenCLの複雑な概念を減らすようにしています.
あと,せっかくなので,なるべくポインタを使わないように書きました.
ポインタもラップしているので,うっかりキャストに失敗するパターンを減らしています.

TinyCLはOpenCLを使う上で事故る心配を少なくしてます.
とにかくこれが大事なので重点的に解決できるよう心がけました(迫真)

14
15
0

Register as a new user and use Qiita more conveniently

  1. You get articles that match your needs
  2. You can efficiently read back useful information
  3. You can use dark theme
What you can do with signing up
14
15