如何在cuda內核函數中產生隨機數（host端調用，device端產生）

最近，需要在kernel函數中調用浮點型的隨機數。於是上網搜了下相關資料，一種方式是自己手動寫一個隨機數的__device__函數，然后在調用的時候調用這個函數。另一種，原來cuda在toolkit中給出了實現方式。

 

首先要用到三個函數：

curandCreateGenerator(&gen,CURAND_RNG_PSEUDO_DEFAULT); 

指定觸發器為gen，隨機方式為CURAND_RNG_PSEUDO_DEFAULT

curandSetPseudoRandomGeneratorSeed(gen,1234ULL);

指定種子為1234ULL(不同的種子產生的隨機數列是不一樣的)

curandGenerateUniform(gen,devData,n);

根據觸發器gen，輸出目標指針為devData，規模大小為n

/*
 * 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;
}

樣例輸出后，產生的結果為：（待補）