Thrust 〜CUDA C++ template library〜

はじめに

先のブログでPyTorch C++を紹介した(その1,その2,その3)。CUDAプログラミングを知らなくても、GPUを利用する深層学習のプログラムをC++で実装できる。ところで、NVIDIAはCUDAを利用するためのC++インターフェースThrustを公開している。これはテンプレートライブラリである。今回はこれを紹介する。

コンテナ

Thrustは2種類のコンテナを提供する。

  • host_vector
  • device_vector
  • 前者はCPU側のメモリ上に、後者はGPU側のメモリ上にバッファを確保する。その使い方は以下の通り。

    thrust::host_vectorstd::vectorとほぼ同じ機能を持つ。前者を使うメリットはデバイス側とのデータの遣り取りがスムーズになることである。

    アルゴリズム

    C++の標準テンプレートライブラリ(Standard Template Library:STL)には、「アルゴリズム」と総称される非常にたくさんの関数が用意されている。その1つにstd::transformがある。使い方は以下の通り。

    上のstd::transformは、svの各値にマイナスを付けdvに代入している。これと同じ機能を持つGPU版は以下の通りである。

    std::transformと同じ名前の関数thrust::transformが用意されているので、これを用いる。ところで、thrust::sequenceはSTLにない関数である。上の引数の場合、1始まりの連続整数値を生成する。

    ここまでに紹介したインターフェースを用いて簡単な画像処理のプログラムを作成し、速度比較を行う。

    画像処理プログラム

    今回実装する画像処理の内容は以下の2つである。

    1. RGB画像をグレイ画像に変換する。
    2. グレイ画像に閾値を与えて2値化する。

    まず最初に、グレイ化に使用する関数オブジェクト(Functor)を示す。

    10行目にグレイ画像への変換式が定義されている。それ以降の行で0から255までの値に収める処理をする。同様に2値化に使用するFunctorは以下のようになる。

    11行目で2値化処理をしている。これらのFunctorが持つメソッドoperator()には識別子「__host__ __device__」が付いている。これをつけることにより、CPU上でもGPU上でも動作する関数になる。この辺りはCUDAプログラミングの流儀である(詳細は割愛する)。これらのFunctorを用いた画像処理プログラムは以下の通り。
    CPU版:

  • 4行目:画像の読み込み。
  • 11行目:画像の中身を配列にコピー。
  • 22行目:グレイ画像化
  • 23行目:2値化
  • 24行目:画像の作成
  • GPU版:

  • 4行目:画像の読み込み。
  • 19,20行目:画像の中身をGPU側のメモリにコピー。
  • 23行目:グレイ画像化
  • 24行目:2値化
  • 25行目:ホスト側(CPU側)へバッファを転送。
  • 26行目:画像の作成
  • それぞれのプログラム内で、forループをITERATIONS(=1000)回まわす。その処理時間を計測すると以下のようになる。

    CPU版 8803 [ms]
    GPU版 103 [ms]

    明らかにGPU版の方が高速である。今回検証に使用したマシーン(EC2インスタンス p2.xlarge)のスペックは以下の通りである。

    CPU Intel Xeon 2.3GHz(x4), CPUメモリ 60GB
    GPU Tesla K80, GPUメモリ 11GB

    まとめ

    今回は、C++のテンプレートライブラリであるThrustを紹介し、高速に処理できることを示した。インターフェースもSTLを意識したものになっており使いやすい。CUDA環境で高速化を実現する際の手っ取り早いツールである。ただし、行列演算ができないため、昨今流行りの深層学習には取り入れ難い。

    Kumada Seiya

    Kumada Seiya

    仕事であろうとなかろうと勉強し続ける、その結果”中身”を知ったエンジニアになれる

    最近の記事

    • 関連記事
    • おすすめ記事
    • 特集記事

    アーカイブ

    カテゴリー

    PAGE TOP