thrustを使ってみた

ここ数日thrustで試しに色々書いていたのでそのことを少しだけ.
thrustというのはGPUで並列計算をするためのライブラリで、だいぶ前からcudaをインストールすると勝手に入ってくれているらしいものです.
device用のポインタがどうこうとかcudaMallocしたらcudaFreeしないといけないとかを考えずにかけるので楽です.
さらに関数名や引数の取り方などがstdと同じ場合が多いので導入が楽かなとか思います.
githubにサンプルコードがたくさんあり、使い方を知るのも結構楽です.

で、どれくらい早くなるのかなーと思いこんなコードを実行してみました.
venctor<int>を8192個とって乱数を入れて10000回インクリメントするのにどれくらい時間がかかるかを求めるだけのプログラムです.

#include<iostream>
#include<algorithm>
#include<iomanip>
#include<time.h>
#include<thrust/host_vector.h>
#include<thrust/device_vector.h>
#include<thrust/sort.h>

#define N (8<<10)
#define C1 10000

template<class T>
class plusOne{
public:
    __device__ __host__ T operator() (T a){
        return a+1;
    }
};

int main(){
    srand(time(NULL));
    thrust::host_vector<int> hv(N);
    std::generate(hv.begin(),hv.end(),rand);
    thrust::device_vector<int> dv=hv;

    cudaEvent_t start,stop;
    float elapsed;
    cudaEventCreate(&start);
    cudaEventCreate(&stop);


    cudaEventRecord(start,0);
    for(int c=0;c<C1;c++){
        thrust::transform(dv.begin(),dv.end(),dv.begin(),plusOne<int>());
    }
    cudaEventRecord(stop,0);
    cudaEventSynchronize(stop);
    cudaEventElapsedTime(&elapsed,start,stop);

    std::cout<<"gpu :"<<elapsed<<"ms ["<<std::setprecision(8)<<C1/elapsed<<"/ms]"<<std::endl;

    std::generate(hv.begin(),hv.end(),rand);
    dv=hv;

    cudaEventRecord(start,0);
    for(int c=0;c<C1;c++){
        thrust::transform(hv.begin(),hv.end(),hv.begin(),plusOne<int>());
    }
    cudaEventRecord(stop,0);
    cudaEventSynchronize(stop);
    cudaEventElapsedTime(&elapsed,start,stop);

    std::cout<<"cpu :"<<elapsed<<"ms ["<<std::setprecision(8)<<C1/elapsed<<"/ms]"<<std::endl;

    cudaEventDestroy(start);
    cudaEventDestroy(stop);

    return 0;
}
で、実行結果がこんな感じ(有効数字は気にしないで)
時間[ms] 計算速度[/ms]
Geforce GTX750Ti 57.9752 172.48764
Xeon E3-1220 v3 13598.045 0.73539984
Tesla K20X 107.292 93.203987
Xeon X5670 16790.818 0.59556359

「わーはやーい」とは思うのですが、直にcudaMallocレベルの関数とかを触っている時ってthreadをいくつにして関数を実行とかしているわけで、thrustは内部でどうやって呼び出しているのかすごく気になりました.

カテゴリー:CUDA
記事作成日:2017-05-30