2016-10-19 9 views
0

デバイスファンクタ内でCURANDとThrustを併用することはできますか?最小コードの例は次のとおりです。推力ファクタ内のCURANDの使用

#include <thrust/device_vector.h> 

struct Move 
{ 
    Move() {} 

    using Position = thrust::tuple<double, double>; 

    __host__ __device__ 
    Position operator()(Position p) 
    { 
     thrust::get<0>(p) += 1.0; // use CURAND to add a random N(0,1) 
     thrust::get<1>(p) += 1.0; // use CURAND to add a random N(0,1) 
     return p; 
    } 
}; 

int main() 
{ 
    // Create vectors on device 
    thrust::device_vector<double> p1(10, 0.0); 
    thrust::device_vector<double> p2(10, 0.0); 

    // Create zip iterators 
    auto pBeg = thrust::make_zip_iterator(thrust::make_tuple(p1.begin(), p2.begin())); 
    auto pEnd = thrust::make_zip_iterator(thrust::make_tuple(p1.end(), p2.end() )); 

    // Move points in the vectors 
    thrust::transform(pBeg, pEnd, pBeg, Move()); 

    // Print result (just for debug) 
    thrust::copy(p1.begin(), p1.end(), std::ostream_iterator<double>(std::cout, "\n")); 
    thrust::copy(p2.begin(), p2.end(), std::ostream_iterator<double>(std::cout, "\n")); 

    return 0; 
} 

演算子関数内で乱数を作成する正しい方法は何ですか?

+0

ここで説明されているcuRANDデバイスAPIを使用する必要があります。http://docs.nvidia.com/cuda/curand/device-api-overview.html#device-api-overview –

答えて

3

デバイスファンクタ内でスラストと一緒にCURANDを使用することはできますか?

はい、可能です。 @ m.sによって示されるように。あなたがcurandから必要とするもののほとんどは、curandのドキュメントのcurand device api exampleから入手できます。 (実際には、推測/徴候のサンプルコードも文書にあります。here

推力アルゴリズムコールで示されているセットアップカーネルの動作を模倣することができます。 thrust::for_each_nを使用して、各デバイスベクトル要素の初期状態変数を設定します。

その後、追加のイテレータを使用して初期化されたcurand状態をMoveファンクタに渡し、ファクタ内でcurand_uniform(たとえば)を呼び出すだけでよい。ここで

は完全にあなたのコードに基づいて、例を働いている:

$ cat t20.cu 
#include <thrust/device_vector.h> 
#include <curand_kernel.h> 
#include <iostream> 
#include <thrust/iterator/counting_iterator.h> 
#include <thrust/transform.h> 
#include <thrust/for_each.h> 

const int seed = 1234; 
const int ds = 10; 
const int offset = 0; 

struct Move 
{ 
    Move() {} 

    using Position = thrust::tuple<double, double, curandState>; 

    __device__ 
    Position operator()(Position p) 
    { 
     curandState s = thrust::get<2>(p); 
     thrust::get<0>(p) += curand_uniform(&s); // use CURAND to add a random N(0,1) 
     thrust::get<1>(p) += curand_uniform(&s); // use CURAND to add a random N(0,1) 
     thrust::get<2>(p) = s; 
     return p; 
    } 
}; 

struct curand_setup 
{ 
    using init_tuple = thrust::tuple<int, curandState &>; 
    __device__ 
    void operator()(init_tuple t){ 
     curandState s; 
     int id = thrust::get<0>(t); 
     curand_init(seed, id, offset, &s); 
     thrust::get<1>(t) = s; 
     } 
}; 

int main() 
{ 
    // Create vectors on device 
    thrust::device_vector<double> p1(ds, 0.0); 
    thrust::device_vector<double> p2(ds, 0.0); 
    thrust::device_vector<curandState> s1(ds); 

    // Create zip iterators 
    auto pBeg = thrust::make_zip_iterator(thrust::make_tuple(p1.begin(), p2.begin(), s1.begin())); 
    auto pEnd = thrust::make_zip_iterator(thrust::make_tuple(p1.end(), p2.end(), s1.end() )); 
    auto pInit = thrust::make_zip_iterator(thrust::make_tuple(thrust::counting_iterator<int>(0), s1.begin())); 
    // initialize random generator 
    thrust::for_each_n(pInit, ds, curand_setup()); 
    // Move points in the vectors 
    thrust::transform(pBeg, pEnd, pBeg, Move()); 

    // Print result (just for debug) 
    thrust::copy(p1.begin(), p1.end(), std::ostream_iterator<double>(std::cout, "\n")); 
    thrust::copy(p2.begin(), p2.end(), std::ostream_iterator<double>(std::cout, "\n")); 

    return 0; 
} 
$ nvcc -arch=sm_61 -std=c++11 t20.cu -o t20 -lcurand 
$ ./t20 
0.145468 
0.820181 
0.550399 
0.29483 
0.914733 
0.868979 
0.321921 
0.782857 
0.0113023 
0.28545 
0.434899 
0.926417 
0.811845 
0.308556 
0.557235 
0.501246 
0.206681 
0.123377 
0.539587 
0.198575 
$ 

この質問について:

オペレータ関数内の乱数を作成するための正しい方法は何ですか?

は推力にcurandを使用しても問題ありませんが、あなたはまた、推力がRNG facilityに建てられており、完全に働いていた使用例hereがあることに注意する必要がある場合があります。

+0

これは完全に機能します。私はCURANDを好んでいます。なぜなら、異なる種の代わりにシーケンスを使って乱数の良質を保証するからです。しかし、 'ds = 1000000'を使ってみたところ、プログラムがクラッシュしました。メモリの問題ではないはずです(curandStateはそれぞれ48bです)。 – GerryR

+0

ds = 1000000で私はクラッシュしません。どんなOS、GPU、CUDAバージョン、GPUアーチをコンパイルしていますか? Windowsのビジュアルスタジオにいる場合は、64ビット(x64)リリースプロジェクトをビルドしていること、win32プロジェクトではなくデバッグプロジェクトでないことを確認してください。ところで、CURANDドキュメント自体が(http://docs.nvidia.com/cuda/curand/device-api-overview.html#thrust-and-curand-example) –

+0

I [スラスト+ CURANDコード例]を有しますUbuntu 16.04、GeForce GTX 745/PCIe/SSE2、CUDA 8.0、arch = sm_50を使用しています: '' thrust :: system :: system_error ''のインスタンスをスローした後に終了するwhat():launched timed – GerryR