Cuda and RNG

Accelerated Computing CUDA CUDA Programming and Performance
1

Our team is trying to find a cuda kernel that will generate us a list of random numbers we can then use in a subsequent kernel. However, I have not been able to figure out the Mersenne Twister in the SDK. Every time we call RandomGPU it screws up our memory. Any ideas on our problem? If anyone knows of any other random number generators that would be awesome too.

Here’s a snippet of where we call the thing.

[codebox]

#define NM_PRTS 1409600 // max number of particles ions or electrons

#define X_GRD 514 // number of x grid points + 1

#define Y_GRD 4098 // number of y grid points + 1

// define system constants

#define PI 3.14159265358979

#define TPI 6.283185307179586476925287

#define ISEED 987654321 // random number seed

#define P0 30.0

#define SCALE 10.0

#define RATO (1.0/400.0)

#define DELT (TPI/64.)

#define BXM 0.0

#define BYM 3.0

#define BZM 0.0

#define SIGMA 10.0 // sigma for cold electrons

#define SIGMA1 10.0 // sigma for cold ions

#define SIGMA2 0.3 // sigma for hot ions

#define SIGMA3 1.0 // sigma for hot electrons

#define TSTART 0.0

#define TMAX 100.0

#define LF 20 // number of iterations between info file outputs

#define LFINT 5 // number of info intervals between output files

#define DX 1.0

#define DY 1.0

// derived parameters

#define NX (X_GRD-1)

#define NX1 (NX-1)

#define NX12 (NX1/2)

#define NY (Y_GRD-1)

#define NY1 (NY-1)

#define NY12 (NY1/2)

// avg number of particles per cell?

#define NIJ 18 // avg number of particle per cell?

…

float *d_Rand;

RandomGPU<<<32, 128>>>(d_Rand, 6NX1NIJ);

…

[/codebox]

2

I don’t know if they’ve changed it since but I used the SDK Mersenne Twister a year ago and found it to be unreliable, i.e. bad numbers. At the moment I’m using a CUDA implementation of the mt19937 which I downloaded from these forums some time back. It carries the NVidia license so I guess it’s from them. I’ve attached the file that contains a bunch of device functions which can be called direct from your kernel. It’s part of a larger package. Google can probably help you find the rest.

I guess discussing the performance of different RNGs is a bit out of the scope of this post. If you PM me I can send you another but it’s not guaranteed to run under Windoze without some tweaking.
mt19937_ref.cu (15.5 KB)