接下来在调用的时候调用这些函数,并且添加CURAND的动态链接库

原稿来自:cuda curand toolkit
document

Translated by xingoo

多年来,要求在kernel函数中调用浮点型的随机数。于是上网搜了下有关材质,一种方法是友善手动写三个自由数的__device__函数,然后在调用的时候调用这么些函数。另一种,原来cuda在toolkit中付出了得以实现格局。

Translated by xingoo

假使有荒唐请联系:xinghl90@gmail.com

 

假设有不当请联系:xinghl90@gmail.com

 

首先要用到七个函数:

 

 

curandCreateGenerator(&gen,CURAND_RNG_PSEUDO_DEFAULT); 

2Host API简述

2.3 返回值

点名触发器为gen,随机方式为CURAND_RNG_PSEUDO_DEFAULT

利用host api,用户需求在头文件的局地含有 curand.h,并且添加CURAND的动态链接库,即在LINKE奥迪Q5的INPUT里面添加”curand.lib;”。那么些文书档案是基于CUDA runtime的,所以用户的代码也应有是在runtime时调用的。而driver API是不帮助CURAND的。

持有的CURAND host端的函数再次回到值都是curandStatus_t.若是调用没有错误,则赶回成功,即再次回到值为CURAND_STATUS_SUCCESS.要是产生了错误,重返值会依照错误的例外而各异。因为cuda允许内核函数异步的调用cpu端的代码,由此回到的一无所能,有恐怕是在调用函数库时发生的,而非CURAND内核函数,此时,重返值是CURAND_STATUS_PREEXISTING.

curandSetPseudoRandomGeneratorSeed(gen,1234ULL);

 

 

钦命种子为1234ULL(分化的种子发生的任性数列是差别的)

触发器生成随机数,CURAND同过内部的包装发生伪随机数列只怕真随机数列。具体的步调如下:

2.4 触发函数

curandGenerateUniform(gen,devData,n);

1 使用函数curandCreateGenerator成立3个新的对象项目(参考 触发器类型)触发器

curandStatus_t curandGenerate(curandGenerator_t generator, unsigned int *outputPtr, size_t num)

遵照触发器gen,输出指标指针为devData,规模大小为n

2 设置触发器的精选(参考 触发器选项),比如,使用curandSetPseudoRandomGeneratorSeed() 来安装种子。

curandGenerate() 函数用来变化伪随机只怕真随机数体系。包含 XO猎豹CS6WOW、M奥迪Q7G32k3a,MTGP32,and SOBOL32,各个成分都以三十八个人unsigned int型的每人都以随意产生的。对于SOBOL64触发器来说,发生的是每一人随机发生的63位的unsigned long long型随机数。

/*
 * This program uses the host CURAND API to generate 100 
 * pseudorandom floats.
 */
#include <stdio.h>
#include <stdlib.h>
#include <cuda.h>
#include <curand.h>

#define CUDA_CALL(x) do { if((x)!=cudaSuccess) { \
    printf("Error at %s:%d\n",__FILE__,__LINE__);\
    return EXIT_FAILURE;}} while(0)
#define CURAND_CALL(x) do { if((x)!=CURAND_STATUS_SUCCESS) { \
    printf("Error at %s:%d\n",__FILE__,__LINE__);\
    return EXIT_FAILURE;}} while(0)

int main(int argc, char *argv[])
{
    size_t n = 100;
    size_t i;
    curandGenerator_t gen;
    float *devData, *hostData;

    /* Allocate n floats on host */
    hostData = (float *)calloc(n, sizeof(float));

    /* Allocate n floats on device */
    CUDA_CALL(cudaMalloc((void **)&devData, n*sizeof(float)));

    /* Create pseudo-random number generator */
    CURAND_CALL(curandCreateGenerator(&gen, 
                CURAND_RNG_PSEUDO_DEFAULT));

    /* Set seed */
    CURAND_CALL(curandSetPseudoRandomGeneratorSeed(gen, 
                1234ULL));

    /* Generate n floats on device */
    CURAND_CALL(curandGenerateUniform(gen, devData, n));

    /* Copy device memory to host */
    CUDA_CALL(cudaMemcpy(hostData, devData, n * sizeof(float),
        cudaMemcpyDeviceToHost));

    /* Show result */
    for(i = 0; i < n; i++) {
        printf("%1.4f ", hostData[i]);
    }
    printf("\n");

    /* Cleanup */
    CURAND_CALL(curandDestroyGenerator(gen));
    CUDA_CALL(cudaFree(devData));
    free(hostData);    
    return EXIT_SUCCESS;
}

3 使用cudaMalloc()申请设备上的贮存空间

curandStatus_t curandGenerateUniform(curandGenerator_t generator,  float *outputPtr, size_t num)

样例输出后,爆发的结果为:(待补)

4 通过curandGenerate()或然其余的生成函数 生成随机数

curandGenerateUniform() 函数用来发生0.0-1.0间的服服帖帖均匀分布(uniformly distributed)的浮点型随机数,当中0不带有0.0,包涵1.0。

 

5 使用生成结果

curandStatus_t curandGenerateNormal(curandGenerator_t generator, float *outputPtr, size_t n, float mean, float stddev)

 

6 一旦指望生成越来越多的轻易数,能够频繁运用curandGenerate().

curandGenerateNormal()函数通过点名的不二法门和输出目的来发生听从正态分布(normally distributed)的浮点型随机数。

7 透过curandDestroyGenertator(),释放地址空间

curandStatus_t curandGenerateLogNormal(curandGenerator_t generator, float *outputPtr, size_t n, float mean, float stddev)

 

curandGenerateLogNormal()函数通过点名的法门和出口指标发出遵循对数正态分布(log-normaly distributed)的浮点数。

即使期待在cpu的host端生成随机数,那么首先步应该调用curandCreateGeneratorHost(),;在第贰步,要申请主机的一段地址空间来接过生成结果。别的步骤,在host端仍旧device端的操作都以同等的。

curandStatus_t curandGeneratePoisson(curandGenerator_t generator, unsigned int *outputPtr, size_t n, double lambda)

 

curandGeneratePoision()函数通过点名的lamda产生基于泊松分布(possion distributed)的随机数。

同一时半刻间段创立多少个触发器是足以的,每一种触发器的包装都独立注脚。每种触发器本人分明一组生成连串。就算老是运营的时候,设置同样的参数,生成随机系列也都以平等的。在device端生成的队列与在host端生成的也是一模一样的。

curandStatus_tcurandGenerateUniformDouble(curandGenerator_t generator, double *outputPtr, size_t num)

 

curandGenerateUniformDouble()函数爆发双精度的均匀分布的随机数。

留意步骤4里面,curandGenerate()调用一段内核函数,并且异步再次来到。所以,假使您在区别的流里面运转了另3个内核函数,而那个内核函数又要运用curandGenerate()的结果,那么就必须调用cudaThreadSynchronized() 来达成共同,以保证生成随机连串的内核函数在被调用前结束执行。(也可以通过下边那种艺术化解:use the stream management/event management routines 实在不领会那一个办法怎么翻译)

curandStatus_tcurandGenerateNormalDouble(curandGenerator_t generator, double *outputPtr, size_t n, double mean, double stddev)

 

curandGenerateNormalDouble()触发器通过点名的措施和规范输出对象发生基高满堂态分布的双精度随机数。双精度的妄动数只可以在总计能力在1.3之上或许host端发生。

透过八个host端的指针来调用在device上运维的触发器,以及通过二个device端的指针来调用运转在host端的触发器,那种情况都属于未定义。

curandStatus_t curandGenerateLogNormalDouble(curandGenerator_t generator,double *outputPtr, size_t n, double mean, double stddev)

 

curandGernerateLogNormalDouble()通过点名的不二法门和输出对象爆发基张成功态分布的对数正态分布双精度随机数。

2.1触发器类型Generator Types

 

随意数触发器通过curandCreateGenerator()来传递类型创制。在CURAND中有7种随机数触发器类型,大家把他们分成两类。

唯有多维度的触发器才能产生真随机数。(差不多是以此意思,For quasirandom generation,the number of results returned must be a multiple of the dimension of the generator)

2.1.1 伪随机数触发器

 

CURAND_RNG_PSEUDO_XOPAJEROWOW,(基于异或位移算法XO帕杰罗WOW)

生成函数可以被同3个触发器数次调用,来发出接二连三的结果块。对于伪随机数触发器,数十次调用size大小的妄动数,相当于一回调用产生n*710官方网站,size大小的随机数。对于真随机数触发器,由于内部存款和储蓄器的空中排序难点,数次短的调用,与贰次长调用发生的结果并不一样;然则,产生的n维动态数组确实相同的。

CURAND_RNG_PSEUDO_MRG32K3A,(基于Combined Multiplie Recursive)

 

CURAND_RNG_PSEUDO_MTGP32(基于Mason旋转法Mersenne Twister)

双精度的任意数只好在估测计算能力1.3上述的装备可能host端发生。

2.1.2 真随机数触发器

 

那4种触发器都是依照SOBOL种类的变种的真随机数系列触发器

2.5 Host API Example

CURAND_RNG_QUASI_SOBOL32(是一种3三人类别的SOBOL触发器)

/*

 * This program uses the host CURAND API to generate 100 

 * pseudorandom floats.

 */

#include <stdio.h>

#include <stdlib.h>

#include <cuda.h>

#include <curand.h>



#define CUDA_CALL(x) do { if((x)!=cudaSuccess) { \

    printf("Error at %s:%d\n",__FILE__,__LINE__);\

    return EXIT_FAILURE;}} while(0)

#define CURAND_CALL(x) do { if((x)!=CURAND_STATUS_SUCCESS) { \

    printf("Error at %s:%d\n",__FILE__,__LINE__);\

    return EXIT_FAILURE;}} while(0)



int main(int argc, char *argv[])

{

    size_t n = 100;

    size_t i;

    curandGenerator_t gen;

float *devData, *hostData;



    /* Allocate n floats on host */

    hostData = (float *)calloc(n, sizeof(float));



    /* Allocate n floats on device */

    CUDA_CALL(cudaMalloc((void **)&devData, n*sizeof(float)));



    /* Create pseudo-random number generator */

    CURAND_CALL(curandCreateGenerator(&gen, 

                CURAND_RNG_PSEUDO_DEFAULT));



    /* Set seed */

    CURAND_CALL(curandSetPseudoRandomGeneratorSeed(gen, 

                1234ULL));



    /* Generate n floats on device */

    CURAND_CALL(curandGenerateUniform(gen, devData, n));



    /* Copy device memory to host */

    CUDA_CALL(cudaMemcpy(hostData, devData, n * sizeof(float),

        cudaMemcpyDeviceToHost));



    /* Show result */

    for(i = 0; i < n; i++) {

        printf("%1.4f ", hostData[i]);

    }

    printf("\n");



    /* Cleanup */

    CURAND_CALL(curandDestroyGenerator(gen));

    CUDA_CALL(cudaFree(devData));

    free(hostData);    

    return EXIT_SUCCESS;

}

CURAND_RNG_QUASI_SCRAMBLED_SOBOL32(srambled不明白怎么翻译才好)

 

CURAND_RNG_QUASI_SOBOL64(是一种六13人体系的SOBOL触发器)

普通经过CURAND库产生的随意数规模越大,发生的本性越好。对于频仍调用发生小范围的随机数来说,尽大概少的调用随机函数库而发生大批量的专断数来使用,更有成效。比如1次发生n*size大小规模的随意数,然后分n次使用,要比n次调用,每一趟产生size高效。XOCRUISERWOW是默许的伪随机数触发器,通过暗中认可的排序,第②回调用可能要开支一些时刻来运营,后来的调用就不须求那步了。为了制止运营时间,能够利用CURAND_ORDERING_PSEUDO_SEEDED排序。

CURAND_RNG_QUASI_SCRAMBLED_SOBOL64

 

 

MTGP32 mersenne Twister算法与线程和块数目紧凑联系。MTGP32的发出结果平常是2个透过一定的参数集发生的钦命顺序的2陆拾七个高低的样例,每陆拾伍个块使用区别的参数集并且每2陆十个线程爆发全体中的多少个样例。因而利用MTGP32产生163八十九个样品是最便捷的。(这一块有待钻探)。

2.2触发器选项 Generator Options

 

开创时,随机数触发器就能通过中央选项—seed种子,offset消耗和order顺序来定义。

 

2.2.1 seed 种子

种子是3个6三人的整型数,用来早先化伪随机触发器的发生。相同的种子,能够发生同样的种类。

2.2.2 Offset消耗/补偿

以此选项用来跳过体系初始的一段随机数。固然offset=100,那么首先个随机数发生的时候,会从原来的任意数类别的第玖0个起来。那让多次运作同一段程序可以几次三番选用同一类别产生的任性数,而不会重叠。这一个跳过的功效不辅助CURAND_RNG_PSEUDO_MTGP32触发器。

2.2.4 Order顺序

那几个选项用来采取在大局内存上生成的妄动数怎么着排序。

有三种伪随机类别顺序能够挑选:

CURAND_ORDERING_PSEUDO_DEFAULT(默认的)

CURAND_ORDERING_PSEUDO_BEST

CURAND_ORDERING_PSEUDO_SEEDED.

有一种真随机类别顺序能够选拔:

CURAND_ORDERING_DEFAULT.

 

目前,CURAND_ORDERING_PSEUDO_DEFAULT 和 CURAND_ORDERING_PSEUDO_BEST对于具有的伪随机触发器产生的结果都是一样的。可是,现在的本子,CURAND大概会调动CURAND_ORDERING_PSEUDO_BEST,使之无论是质量照旧品质上都足以改进。而CURAND_ORDERING_PSEUDO_DEFAULT对于其余的CURAND版本都会维持同样的产生结果。在近年来的版本中,只有XOLANDWOW触发器拥有两种发生顺序。

 

有关发生顺序,每个触发器的做法如下:

XO冠道WOW伪随机触发器:

CURAND_ORDERING_PSEUDO_BEST

  近来版本同CURAND_ORDERING_PSEUDO_DEFAULT的出口顺序是均等的

CURAND_ORDERING_PSEUDO_DEFAULT

  当offset=n时,随机系列在全局内部存款和储蓄器中从(n mpd 4096)-2 67+[n/4096]开始

CURAND_ORDERING_PSEUDO_SEEDED

  当offset=n时,随机体系在大局内部存储器中从n mod 4096初阶。也正是说,每409五个线程使用不一致的种子。那各样子的主意,纵然节省了运维时刻,但对于一些种子的值,伪随机连串输出结果总计也出示出了部分不足。

 

M瑞鹰G32K3A 伪随机触发器

CURAND_ORDERING_PSEUDO_BEST

  跟上面包车型地铁DEFAULT一样

CURAND_ORDERING_PSEUDO_DEFAULT

  当offset=n时,公式为(n mod 4096) -276+[n/4096](注意那一个触发器发生的跨度跟XO揽胜WOW产生的跨度是不雷同的)

 

MTGP32 伪随机触发器

CURAND_ORDERING_PSEUDO_BEST

  同样跟DEFAULT的一一相同

CURAND_ORDERING_PSEUDO_DEFAULT

  MTGP32触发器基于基本算法,通过设置分歧参数,可以产生64种随机种类。使用S(p)来定义种类,p为参数。Offset=n时,公式为S([n/256]mod 64).也正是说,S(0)会有2六二十一个例子,然后是S(1)的257个,一向到S(63),一向重复。

 

Philox_4x32_10 伪随机触发器

CURAND_ORDERING_PSEUDO_BEST

  跟上面包车型地铁DEFAULT一样

CURAND_ORDERING_PSEUDO_DEFAULT

  种种触发器的线程都会时有发生区别的人身自由连串。在host API中有8192种不一样的行列。来自于贰个行列的每几个值都重视于下二个队列的多个值(Each four values from one sequence are followed by four values from next sequence)。

 

三13位和6三位的SOBOL以及赶快SOBOL真随机触发器

CURAND_ORDERING_QUASI_DEFAULT

  当在d个dimensions上边生成n个结果,第3个dimesion由n/d个结果组成,第一个也由n/d个组成,平素到第d个dimension.触发器只好生成dimension的倍数大小的体系。这一个dimension的参数d使用curandSetQuasiRandomGeneratorDimensions(),并且暗许为1.

 

 

 

 

相关文章