pseudorandom number(무작위 난수) 생성을 위한 NVIDIA cuRAND 라이브러리 - 1
cuRAND 라이브러리 :
pseudorandom number(무작위 난수) 와 quasirandom number(준 난수) 생성을 위한 라이브러리
cuRAND 라이브러는 library on the host(CPU)와 device(GPU) 헤더 파일 두 가지로 구성
1. host에서 라이브러리 함수 호출 시 device global memory에 난수가 저장되고, 이를 host memory로 복사해 와서 사용할 수 있다.
2. device 헤더 파일은 random number generator(난수 생성기) state 세팅하기 위한 device 함수와, sequences of random numbers를 생성하는 것이 정의됨.
즉, host code에서 host library를 사용해서 device에서 생성된 난수를 가져와 사용하거나 (1),
아니면 사용자 커널 작성 시 커널에서 난수를 사용하기 위한 것 차이 같다(2)...
>> (2) 사용 시 사용자 커널에서 device global memory에 쓰고 읽는 것이 아니기 때문에 빠를 것으로 예상
PS. This allows random numbers to be generated and immediately consumed by user kernels without requiring the random numbers to be written to and then read from global memory.
Host API Overview
#include "curand.h" >> host code에서 난수 생성에 필요한 함수들을 호출 하기 위해
난수 생성을 위한 generator가 필요 >> 난수 생성을 위해 필요한 모든 internal state를 가짐.
일반적 난수 생성 과정
1. generator 생성 >> curandCreateGenerator()
- 9개의 generator가 있다.
curandStatus_t CURANDAPI
curandCreateGenerator(curandGenerator_t *generator, curandRngType_t rng_type);
typedef enum curandRngType curandRngType_t;
/**
* CURAND generator types
*/
enum curandRngType {
CURAND_RNG_TEST = 0,
CURAND_RNG_PSEUDO_DEFAULT = 100, ///< Default pseudorandom generator
CURAND_RNG_PSEUDO_XORWOW = 101, ///< XORWOW pseudorandom generator
CURAND_RNG_PSEUDO_MRG32K3A = 121, ///< MRG32k3a pseudorandom generator
CURAND_RNG_PSEUDO_MTGP32 = 141, ///< Mersenne Twister MTGP32 pseudorandom generator
CURAND_RNG_PSEUDO_MT19937 = 142, ///< Mersenne Twister MT19937 pseudorandom generator
CURAND_RNG_PSEUDO_PHILOX4_32_10 = 161, ///< PHILOX-4x32-10 pseudorandom generator
CURAND_RNG_QUASI_DEFAULT = 200, ///< Default quasirandom generator
CURAND_RNG_QUASI_SOBOL32 = 201, ///< Sobol32 quasirandom generator
CURAND_RNG_QUASI_SCRAMBLED_SOBOL32 = 202, ///< Scrambled Sobol32 quasirandom generator
CURAND_RNG_QUASI_SOBOL64 = 203, ///< Sobol64 quasirandom generator
CURAND_RNG_QUASI_SCRAMBLED_SOBOL64 = 204 ///< Scrambled Sobol64 quasirandom generator
};
// 생성 예
curandGenerator_t gen;
curnadCreateGenerator(&gen, CURAND_RNG_PSEUDO_DEFAULT);
2. generator option 세팅 >> curandSetPseudoRandomGeneratorSeed()... etc
>> generator 생성 후 seed, offset, order 옵션을 사용하여 정의될 수 있다
- seed : 64-bit interger 타입. 의사 난수 생성기의 starting state initialize. 같은 seed 값은 항상 같은 결과를 낸다.
- offset : The offset parameter is used to skip ahead in the sequence. If offset = 100, the first random number generated will be the 100th in the sequence. This allows multiple runs of the same program to continue generating results from the same sequence without overlap.
> 난수 생성 seqeunce에서 offset 만큼 건너 뜀. 같은 프로그램을 여러번 실행시 같은 난수 sequence로 부터 중복 없이 결과를 생성할 수 있다(?).
- order : device global memory에 어떤 순서(자세한 설명 찾지 못함)로 저장할 것인가 >> cuRAND 퍼포먼스에 영향을 미친다.
의사난수 경우 기본 값 : CURAND_ORERING_PSEUDO_DEFALUT.
curandStatus_t CURANDAPI
curandSetPseudoRandomGeneratorSeed(curandGenerator_t generator, unsigned long long seed);
curandStatus_t CURANDAPI
curandSetGeneratorOffset(curandGenerator_t generator, unsigned long long offset);
curandStatus_t CURANDAPI
curandSetGeneratorOrdering(curandGenerator_t generator, curandOrdering_t order);
3. 난수 저장을 위한 device memory 할당 >> cudaMalloc()
4. 난수 생성 >> curandGenerate()
예) curandGenerateUniform(생성기, 디바이스 메모리 포인터, 사이즈);
Generation Funcrtion
1. curandGenerate() : 32-bit unsigned interger 반환.
2. curandGenerateUniform() : 균등 분포 float 반환 0 ~ 1 사이. 0은 포함x 1은 포함된다.
- curandStatus_t curandGenerateUniform(curandGenerator_t generator, float *outputPtr, size_t num)
3. curnadGenerateUniformDouble() :균등 분포 double 반환
등등 여러개 존재
5. 난수 생성 값 host로 복사 후 사용
6. generator 파괴 >> curandDestroyGenerator()
The sequence of numbers produced by each generator is deterministic. Given the same set-up parameters, the same sequence will be generated with every run of the program.
- 동일한 generator를 사용 및 세팅 값을 동일하게 하면 항상 같은 난수가 생성
모든 cuRAND host 라이브러리는 curandStatus_t 를 반환 >> 예외 처리 시 사용 위해서. 에러 없이 실행되면 CURAND_STATUS_SUCCESS 반환. 에러 시 이것과 다른 값 반환
난수 생성 예제
/*
* This program uses the host CURAND API to generate 100
* pseudorandom floats.
*/
#include <stdio.h>
#include <stdlib.h>
#include <cuda_runtime.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; // 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;
}
PS. curandGenerate() 커널은 비 동기적으로 실행 >> 그래서 다른 스트림에서 curandGenerate() 결과 값을 사용하려면 동기화가 필요.