QUDA  0.9.0
random_quda.h
Go to the documentation of this file.
1 
2 #ifndef RANDOM_GPU_H
3 #define RANDOM_GPU_H
4 
5 //#include <stdio.h>
6 //#include <string.h>
7 //#include <iostream>
8 #include <curand_kernel.h>
9 
10 namespace quda {
11 
12 #if defined(XORWOW)
13 typedef struct curandStateXORWOW cuRNGState;
14 #elif defined(MRG32k3a)
15 typedef struct curandStateMRG32k3a cuRNGState;
16 #else
17 typedef struct curandStateMRG32k3a cuRNGState;
18 #endif
19 
23 class RNG {
24 public:
25  RNG(int rng_sizes, int seedin, const int XX[4]);
26  RNG(int rng_sizes, int seedin);
28  void Release();
30  void Init();
32  int Size(){ return rng_size;};
33  int Node_Offset(){ return node_offset;};
34  int Seed(){ return seed;};
35  __host__ __device__ __inline__ cuRNGState* State(){ return state;};
37  void restore();
39  void backup();
40 private:
46  int seed;
48  int rng_size;
51  int X[4];
53  void AllocateRNG();
55  void INITRNG(int rng_sizes, int seedin, int offsetin);
56 };
57 
58 
59 
60 
61 
69 template<class Real>
70 inline __device__ Real Random(cuRNGState &state, Real a, Real b){
71  Real res;
72  return res;
73 }
74 
75 template<>
76 inline __device__ float Random<float>(cuRNGState &state, float a, float b){
77  return a + (b - a) * curand_uniform(&state);
78 }
79 
80 template<>
81 inline __device__ double Random<double>(cuRNGState &state, double a, double b){
82  return a + (b - a) * curand_uniform_double(&state);
83 }
84 
90 template<class Real>
91 inline __device__ Real Random(cuRNGState &state){
92  Real res;
93  return res;
94 }
95 
96 template<>
97 inline __device__ float Random<float>(cuRNGState &state){
98  return curand_uniform(&state);
99 }
100 
101 template<>
102 inline __device__ double Random<double>(cuRNGState &state){
103  return curand_uniform_double(&state);
104 }
105 
106 
107 template<class Real>
108 struct uniform { };
109 template<>
110 struct uniform<float> {
111  __device__
112  static inline float rand(cuRNGState &state) {
113  return curand_uniform(&state);
114  }
115 };
116 template<>
117 struct uniform<double> {
118  __device__
119  static inline double rand(cuRNGState &state) {
120  return curand_uniform_double(&state);
121  }
122 };
123 
124 
125 
126 template<class Real>
127 struct normal { };
128 template<>
129 struct normal<float> {
130  __device__
131  static inline float rand(cuRNGState &state) {
132  return curand_normal(&state);
133  }
134 };
135 template<>
136 struct normal<double> {
137  __device__
138  static inline double rand(cuRNGState &state) {
139  return curand_normal_double(&state);
140  }
141 };
142 
143 
144 }
145 
146 #endif
void AllocateRNG()
allocate curand rng states array in device memory
Definition: random.cu:155
static __device__ double rand(cuRNGState &state)
Definition: random_quda.h:138
void Init()
Initialize CURAND RNG states.
Definition: random.cu:146
struct curandStateMRG32k3a cuRNGState
Definition: random_quda.h:17
RNG(int rng_sizes, int seedin, const int XX[4])
Definition: random.cu:122
__device__ Real Random(cuRNGState &state, Real a, Real b)
Return a random number between a and b.
Definition: random_quda.h:70
static __device__ double rand(cuRNGState &state)
Definition: random_quda.h:119
void backup()
Backup CURAND array states initialization.
Definition: random.cu:189
int Size()
return curand rng array size
Definition: random_quda.h:32
cuRNGState * state
Definition: random_quda.h:42
#define b
static __device__ float rand(cuRNGState &state)
Definition: random_quda.h:112
int Seed()
Definition: random_quda.h:34
void Release()
Release Device memory for CURAND RNG states.
Definition: random.cu:168
void restore()
Restore CURAND array states initialization.
Definition: random.cu:179
__device__ float Random< float >(cuRNGState &state, float a, float b)
Definition: random_quda.h:76
Class declaration to initialize and hold CURAND RNG states.
Definition: random_quda.h:23
int rng_size
number of curand states
Definition: random_quda.h:48
void INITRNG(int rng_sizes, int seedin, int offsetin)
CURAND array states initialization.
cuRNGState * backup_state
Definition: random_quda.h:44
int Node_Offset()
Definition: random_quda.h:33
static __device__ float rand(cuRNGState &state)
Definition: random_quda.h:131
__host__ __device__ __inline__ cuRNGState * State()
Definition: random_quda.h:35
__device__ double Random< double >(cuRNGState &state, double a, double b)
Definition: random_quda.h:81
int node_offset
offset in the index, in case of multigpus
Definition: random_quda.h:50
#define a
int X[4]
Definition: random_quda.h:51