ESBMC
Efficient SMT-based Context-Bounded Model Checker
Loading...
Searching...
No Matches
curand_kernel.h File Reference
#include "curand_precalc.h"
#include <cmath>
#include <curand.h>
Include dependency graph for curand_kernel.h:
This graph shows which files directly or indirectly include this file:

Go to the source code of this file.

Classes

struct  curandStateTest
struct  curandStateXORWOW
struct  curandStateSobol32
struct  curandStateScrambledSobol32
struct  curandStateSobol64
struct  curandStateScrambledSobol64

Macros

#define MAX_XOR_N   (5)
#define SKIPAHEAD_BLOCKSIZE   (4)
#define SKIPAHEAD_MASK   ((1 << SKIPAHEAD_BLOCKSIZE) - 1)
#define CURAND_2POW32_INV   (2.3283064e-10f)
#define CURAND_2POW32_INV_DOUBLE   (2.3283064365386963e-10)
#define CURAND_2POW53_INV_DOUBLE   (1.1102230246251565e-16)
#define CURAND_2POW32_INV_2PI   (2.3283064e-10f * 6.2831855f)
#define CURAND_2POW53_INV_2PI_DOUBLE    (1.1102230246251565e-16 * 6.2831853071795860)
#define CURAND_SQRT2   (-1.4142135f)
#define CURAND_SQRT2_DOUBLE   (-1.4142135623730951)
#define PRECALC_NUM_MATRICES   (8)
#define PRECALC_BLOCK_SIZE   (2)
#define PRECALC_BLOCK_MASK   ((1 << PRECALC_BLOCK_SIZE) - 1)
#define XORWOW_SEQUENCE_SPACING   (67)
#define QUALIFIERS   static inline __device__

Typedefs

typedef struct curandStateTest curandStateTest_t
typedef struct curandStateXORWOW curandStateXORWOW_t
 CURAND XORWOW state.
typedef struct curandStateSobol32 curandStateSobol32_t
 CURAND Sobol32 state.
typedef struct curandStateScrambledSobol32 curandStateScrambledSobol32_t
 CURAND Scrambled Sobol32 state.
typedef struct curandStateSobol64 curandStateSobol64_t
 CURAND Sobol64 state.
typedef struct curandStateScrambledSobol64 curandStateScrambledSobol64_t
 CURAND Scrambled Sobol64 state.
typedef struct curandStateXORWOW curandState_t
 Default RNG.
typedef struct curandStateXORWOW curandState

Functions

void __curand_matvec (unsigned int *vector, unsigned int *matrix, unsigned int *result, int n)
void __curand_matidentity (unsigned int *matrix, int n)
void __curand_matmat (unsigned int *matrixA, unsigned int *matrixB, int n)
void __curand_veccopy (unsigned int *vector, unsigned int *vectorA, int n)
void __curand_matcopy (unsigned int *matrix, unsigned int *matrixA, int n)
void __curand_matpow (unsigned int *matrix, unsigned int *matrixA, unsigned long long p, int n)
float __curand_uint32AsFloat (unsigned int i)
double __curand_hilouint32AsDouble (unsigned int hi, unsigned int lo)
float __curand_uint32_as_float (unsigned int x)
void curand_init (unsigned long long seed, unsigned long long subsequence, unsigned long long offset, curandStateTest_t *state)
unsigned int curand (curandStateTest_t *state)
void skipahead (unsigned long long n, curandStateTest_t *state)
template<typename T, int n>
void __curand_generate_skipahead_matrix_xor (unsigned int matrix[])
template<typename T, int n>
void _skipahead_scratch (unsigned long long x, T *state, unsigned int *scratch)
template<typename T, int n>
void _skipahead_sequence_scratch (unsigned long long x, T *state, unsigned int *scratch)
void skipahead (unsigned long long n, curandStateXORWOW_t *state)
 Update XORWOW state to skip n elements.
void skipahead_sequence (unsigned long long n, curandStateXORWOW_t *state)
 Update XORWOW state to skip ahead n subsequences.
void _curand_init_scratch (unsigned long long seed, unsigned long long subsequence, unsigned long long offset, curandStateXORWOW_t *state, unsigned int *scratch)
void curand_init (unsigned long long seed, unsigned long long subsequence, unsigned long long offset, curandStateXORWOW_t *state)
 Initialize XORWOW state.
unsigned int curand (curandStateXORWOW_t *state)
 Return 32-bits of pseudorandomness from an XORWOW generator.
template<typename T>
void skipahead (unsigned int n, T state)
 Update Sobol32 state to skip n elements.
template<typename T>
void skipahead (unsigned long long n, T state)
 Update Sobol64 state to skip n elements.
template<typename XT>
int __curand_find_trailing_zero (XT x)
 Initialize Sobol32 state.
void curand_init (curandDirectionVectors64_t direction_vectors, unsigned long long offset, curandStateSobol64_t *state)
 Initialize Sobol64 state.
template<typename PT>
void _skipahead_stride (int n_log2, PT state)
void curand_init (curandDirectionVectors64_t direction_vectors, unsigned long long scramble_c, unsigned long long offset, curandStateScrambledSobol64_t *state)
 Initialize Scrambled Sobol64 state.
unsigned int curand (curandStateSobol32_t *state)
 Return 32-bits of quasirandomness from a Sobol32 generator.
unsigned int curand (curandStateScrambledSobol32_t *state)
 Return 32-bits of quasirandomness from a scrambled Sobol32 generator.
unsigned long long curand (curandStateSobol64_t *state)
 Return 64-bits of quasirandomness from a Sobol64 generator.
unsigned long long curand (curandStateScrambledSobol64_t *state)
 Return 64-bits of quasirandomness from a scrambled Sobol64 generator.
float _curand_uniform (unsigned int x)
float _curand_uniform (unsigned long long x)
double _curand_uniform_double (unsigned int x)
double _curand_uniform_double (unsigned long long x)
double _curand_uniform_double_hq (unsigned int x, unsigned int y)
double _curand_uniform_double_64 (unsigned long long x)
float curand_uniform (curandStateTest_t *state)
double curand_uniform_double (curandStateTest_t *state)
float curand_uniform (curandStateXORWOW_t *state)
 Return a uniformly distributed float from an XORWOW generator.
double curand_uniform_double (curandStateXORWOW_t *state)
 Return a uniformly distributed double from an XORWOW generator.
float curand_uniform (curandStateSobol32_t *state)
 Return a uniformly distributed float from a Sobol32 generator.
double curand_uniform_double (curandStateSobol32_t *state)
 Return a uniformly distributed double from a Sobol32 generator.
float curand_uniform (curandStateScrambledSobol32_t *state)
 Return a uniformly distributed float from a scrambled Sobol32 generator.
double curand_uniform_double (curandStateScrambledSobol32_t *state)
 Return a uniformly distributed double from a scrambled Sobol32 generator.
float curand_uniform (curandStateSobol64_t *state)
 Return a uniformly distributed float from a Sobol64 generator.
double curand_uniform_double (curandStateSobol64_t *state)
 Return a uniformly distributed double from a Sobol64 generator.
float curand_uniform (curandStateScrambledSobol64_t *state)
 Return a uniformly distributed float from a scrambled Sobol64 generator.
double curand_uniform_double (curandStateScrambledSobol64_t *state)
 Return a uniformly distributed double from a scrambled Sobol64 generator.

Variables

static unsigned int precalc_xorwow_matrix [8][2] = {{850664906UL, 4258393217UL}}

Macro Definition Documentation

◆ CURAND_2POW32_INV

#define CURAND_2POW32_INV   (2.3283064e-10f)

◆ CURAND_2POW32_INV_2PI

#define CURAND_2POW32_INV_2PI   (2.3283064e-10f * 6.2831855f)

◆ CURAND_2POW32_INV_DOUBLE

#define CURAND_2POW32_INV_DOUBLE   (2.3283064365386963e-10)

◆ CURAND_2POW53_INV_2PI_DOUBLE

#define CURAND_2POW53_INV_2PI_DOUBLE    (1.1102230246251565e-16 * 6.2831853071795860)

◆ CURAND_2POW53_INV_DOUBLE

#define CURAND_2POW53_INV_DOUBLE   (1.1102230246251565e-16)

◆ CURAND_SQRT2

#define CURAND_SQRT2   (-1.4142135f)

◆ CURAND_SQRT2_DOUBLE

#define CURAND_SQRT2_DOUBLE   (-1.4142135623730951)

◆ MAX_XOR_N

#define MAX_XOR_N   (5)

◆ PRECALC_BLOCK_MASK

#define PRECALC_BLOCK_MASK   ((1 << PRECALC_BLOCK_SIZE) - 1)

◆ PRECALC_BLOCK_SIZE

#define PRECALC_BLOCK_SIZE   (2)

◆ PRECALC_NUM_MATRICES

#define PRECALC_NUM_MATRICES   (8)

◆ QUALIFIERS

#define QUALIFIERS   static inline __device__

◆ SKIPAHEAD_BLOCKSIZE

#define SKIPAHEAD_BLOCKSIZE   (4)

◆ SKIPAHEAD_MASK

#define SKIPAHEAD_MASK   ((1 << SKIPAHEAD_BLOCKSIZE) - 1)

◆ XORWOW_SEQUENCE_SPACING

#define XORWOW_SEQUENCE_SPACING   (67)

Typedef Documentation

◆ curandState

◆ curandState_t

Default RNG.

◆ curandStateScrambledSobol32_t

CURAND Scrambled Sobol32 state.

◆ curandStateScrambledSobol64_t

CURAND Scrambled Sobol64 state.

◆ curandStateSobol32_t

CURAND Sobol32 state.

◆ curandStateSobol64_t

CURAND Sobol64 state.

◆ curandStateTest_t

◆ curandStateXORWOW_t

CURAND XORWOW state.

Function Documentation

◆ __curand_find_trailing_zero()

template<typename XT>
int __curand_find_trailing_zero ( XT x)

Initialize Sobol32 state.

Initialize Sobol32 state in state with the given direction vectors and offset.

The direction vector is a device pointer to an array of 32 unsigned ints. All input values of offset are legal.

Parameters
direction_vectors- Pointer to array of 32 unsigned ints representing the direction vectors for the desired dimension
offset- Absolute offset into sequence
state- Pointer to state to initialize

Initialize Scrambled Sobol32 state.

Initialize Sobol32 state in state with the given direction vectors and offset.

The direction vector is a device pointer to an array of 32 unsigned ints. All input values of offset are legal.

Parameters
direction_vectors- Pointer to array of 32 unsigned ints representing the direction vectors for the desired dimension
scramble_cScramble constant
offset- Absolute offset into sequence
state- Pointer to state to initialize

◆ __curand_generate_skipahead_matrix_xor()

template<typename T, int n>
void __curand_generate_skipahead_matrix_xor ( unsigned int matrix[])

◆ __curand_hilouint32AsDouble()

double __curand_hilouint32AsDouble ( unsigned int hi,
unsigned int lo )

◆ __curand_matcopy()

void __curand_matcopy ( unsigned int * matrix,
unsigned int * matrixA,
int n )

◆ __curand_matidentity()

void __curand_matidentity ( unsigned int * matrix,
int n )

◆ __curand_matmat()

void __curand_matmat ( unsigned int * matrixA,
unsigned int * matrixB,
int n )

◆ __curand_matpow()

void __curand_matpow ( unsigned int * matrix,
unsigned int * matrixA,
unsigned long long p,
int n )

◆ __curand_matvec()

void __curand_matvec ( unsigned int * vector,
unsigned int * matrix,
unsigned int * result,
int n )

◆ __curand_uint32_as_float()

float __curand_uint32_as_float ( unsigned int x)

◆ __curand_uint32AsFloat()

float __curand_uint32AsFloat ( unsigned int i)

◆ __curand_veccopy()

void __curand_veccopy ( unsigned int * vector,
unsigned int * vectorA,
int n )

◆ _curand_init_scratch()

void _curand_init_scratch ( unsigned long long seed,
unsigned long long subsequence,
unsigned long long offset,
curandStateXORWOW_t * state,
unsigned int * scratch )

◆ _curand_uniform() [1/2]

float _curand_uniform ( unsigned int x)

◆ _curand_uniform() [2/2]

float _curand_uniform ( unsigned long long x)

◆ _curand_uniform_double() [1/2]

double _curand_uniform_double ( unsigned int x)

◆ _curand_uniform_double() [2/2]

double _curand_uniform_double ( unsigned long long x)

◆ _curand_uniform_double_64()

double _curand_uniform_double_64 ( unsigned long long x)

◆ _curand_uniform_double_hq()

double _curand_uniform_double_hq ( unsigned int x,
unsigned int y )

◆ _skipahead_scratch()

template<typename T, int n>
void _skipahead_scratch ( unsigned long long x,
T * state,
unsigned int * scratch )

◆ _skipahead_sequence_scratch()

template<typename T, int n>
void _skipahead_sequence_scratch ( unsigned long long x,
T * state,
unsigned int * scratch )

◆ _skipahead_stride()

template<typename PT>
void _skipahead_stride ( int n_log2,
PT state )

◆ curand() [1/6]

unsigned int curand ( curandStateScrambledSobol32_t * state)

Return 32-bits of quasirandomness from a scrambled Sobol32 generator.

Return 32-bits of quasirandomness from the scrambled Sobol32 generator in state, increment position of generator by one.

Parameters
state- Pointer to state to update
Returns
32-bits of quasirandomness as an unsigned int, all bits valid to use.

◆ curand() [2/6]

unsigned long long curand ( curandStateScrambledSobol64_t * state)

Return 64-bits of quasirandomness from a scrambled Sobol64 generator.

Return 64-bits of quasirandomness from the scrambled Sobol32 generator in state, increment position of generator by one.

Parameters
state- Pointer to state to update
Returns
64-bits of quasirandomness as an unsigned long long, all bits valid to use.

◆ curand() [3/6]

unsigned int curand ( curandStateSobol32_t * state)

Return 32-bits of quasirandomness from a Sobol32 generator.

Return 32-bits of quasirandomness from the Sobol32 generator in state, increment position of generator by one.

Parameters
state- Pointer to state to update
Returns
32-bits of quasirandomness as an unsigned int, all bits valid to use.

◆ curand() [4/6]

unsigned long long curand ( curandStateSobol64_t * state)

Return 64-bits of quasirandomness from a Sobol64 generator.

Return 64-bits of quasirandomness from the Sobol64 generator in state, increment position of generator by one.

Parameters
state- Pointer to state to update
Returns
64-bits of quasirandomness as an unsigned long long, all bits valid to use.

◆ curand() [5/6]

unsigned int curand ( curandStateTest_t * state)

◆ curand() [6/6]

unsigned int curand ( curandStateXORWOW_t * state)

Return 32-bits of pseudorandomness from an XORWOW generator.

Return 32-bits of pseudorandomness from the XORWOW generator in state, increment position of generator by one.

Parameters
state- Pointer to state to update
Returns
32-bits of pseudorandomness as an unsigned int, all bits valid to use.

◆ curand_init() [1/4]

void curand_init ( curandDirectionVectors64_t direction_vectors,
unsigned long long offset,
curandStateSobol64_t * state )

Initialize Sobol64 state.

Initialize Sobol64 state in state with the given direction vectors and offset.

The direction vector is a device pointer to an array of 64 unsigned long longs. All input values of offset are legal.

Parameters
direction_vectors- Pointer to array of 64 unsigned long longs representing the direction vectors for the desired dimension
offset- Absolute offset into sequence
state- Pointer to state to initialize

◆ curand_init() [2/4]

void curand_init ( curandDirectionVectors64_t direction_vectors,
unsigned long long scramble_c,
unsigned long long offset,
curandStateScrambledSobol64_t * state )

Initialize Scrambled Sobol64 state.

Initialize Sobol64 state in state with the given direction vectors and offset.

The direction vector is a device pointer to an array of 64 unsigned long longs. All input values of offset are legal.

Parameters
direction_vectors- Pointer to array of 64 unsigned long longs representing the direction vectors for the desired dimension
scramble_cScramble constant
offset- Absolute offset into sequence
state- Pointer to state to initialize

◆ curand_init() [3/4]

void curand_init ( unsigned long long seed,
unsigned long long subsequence,
unsigned long long offset,
curandStateTest_t * state )

◆ curand_init() [4/4]

void curand_init ( unsigned long long seed,
unsigned long long subsequence,
unsigned long long offset,
curandStateXORWOW_t * state )

Initialize XORWOW state.

Initialize XORWOW state in state with the given seed, subsequence, and offset.

All input values of seed, subsequence, and offset are legal. Large values for subsequence and offset require more computation and so will take more time to complete.

A value of 0 for seed sets the state to the values of the original published version of the xorwow algorithm.

Parameters
seed- Arbitrary bits to use as a seed
subsequence- Subsequence to start at
offset- Absolute offset into sequence
state- Pointer to state to initialize

◆ curand_uniform() [1/6]

float curand_uniform ( curandStateScrambledSobol32_t * state)

Return a uniformly distributed float from a scrambled Sobol32 generator.

Return a uniformly distributed float between 0.0f and 1.0f from the scrambled Sobol32 generator in state, increment position of generator. Output range excludes 0.0f but includes 1.0f. Denormalized floating point outputs are never returned.

The implementation is guaranteed to use a single call to curand().

Parameters
state- Pointer to state to update
Returns
uniformly distributed float between 0.0f and 1.0f

◆ curand_uniform() [2/6]

float curand_uniform ( curandStateScrambledSobol64_t * state)

Return a uniformly distributed float from a scrambled Sobol64 generator.

Return a uniformly distributed float between 0.0f and 1.0f from the scrambled Sobol64 generator in state, increment position of generator. Output range excludes 0.0f but includes 1.0f. Denormalized floating point outputs are never returned.

The implementation is guaranteed to use a single call to curand().

Parameters
state- Pointer to state to update
Returns
uniformly distributed float between 0.0f and 1.0f

◆ curand_uniform() [3/6]

float curand_uniform ( curandStateSobol32_t * state)

Return a uniformly distributed float from a Sobol32 generator.

Return a uniformly distributed float between 0.0f and 1.0f from the Sobol32 generator in state, increment position of generator. Output range excludes 0.0f but includes 1.0f. Denormalized floating point outputs are never returned.

The implementation is guaranteed to use a single call to curand().

Parameters
state- Pointer to state to update
Returns
uniformly distributed float between 0.0f and 1.0f

◆ curand_uniform() [4/6]

float curand_uniform ( curandStateSobol64_t * state)

Return a uniformly distributed float from a Sobol64 generator.

Return a uniformly distributed float between 0.0f and 1.0f from the Sobol64 generator in state, increment position of generator. Output range excludes 0.0f but includes 1.0f. Denormalized floating point outputs are never returned.

The implementation is guaranteed to use a single call to curand().

Parameters
state- Pointer to state to update
Returns
uniformly distributed float between 0.0f and 1.0f

◆ curand_uniform() [5/6]

float curand_uniform ( curandStateTest_t * state)

◆ curand_uniform() [6/6]

float curand_uniform ( curandStateXORWOW_t * state)

Return a uniformly distributed float from an XORWOW generator.

Return a uniformly distributed float between 0.0f and 1.0f from the XORWOW generator in state, increment position of generator. Output range excludes 0.0f but includes 1.0f. Denormalized floating point outputs are never returned.

The implementation may use any number of calls to curand() to get enough random bits to create the return value. The current implementation uses one call.

Parameters
state- Pointer to state to update
Returns
uniformly distributed float between 0.0f and 1.0f

◆ curand_uniform_double() [1/6]

double curand_uniform_double ( curandStateScrambledSobol32_t * state)

Return a uniformly distributed double from a scrambled Sobol32 generator.

Return a uniformly distributed double between 0.0 and 1.0 from the scrambled Sobol32 generator in state, increment position of generator. Output range excludes 0.0 but includes 1.0. Denormalized floating point outputs are never returned.

The implementation is guaranteed to use a single call to curand() to preserve the quasirandom properties of the sequence.

Parameters
state- Pointer to state to update
Returns
uniformly distributed double between 0.0 and 1.0

◆ curand_uniform_double() [2/6]

double curand_uniform_double ( curandStateScrambledSobol64_t * state)

Return a uniformly distributed double from a scrambled Sobol64 generator.

Return a uniformly distributed double between 0.0 and 1.0 from the scrambled Sobol64 generator in state, increment position of generator. Output range excludes 0.0 but includes 1.0. Denormalized floating point outputs are never returned.

The implementation is guaranteed to use a single call to curand() to preserve the quasirandom properties of the sequence.

Parameters
state- Pointer to state to update
Returns
uniformly distributed double between 0.0 and 1.0

◆ curand_uniform_double() [3/6]

double curand_uniform_double ( curandStateSobol32_t * state)

Return a uniformly distributed double from a Sobol32 generator.

Return a uniformly distributed double between 0.0 and 1.0 from the Sobol32 generator in state, increment position of generator. Output range excludes 0.0 but includes 1.0. Denormalized floating point outputs are never returned.

The implementation is guaranteed to use a single call to curand() to preserve the quasirandom properties of the sequence.

Parameters
state- Pointer to state to update
Returns
uniformly distributed double between 0.0 and 1.0

◆ curand_uniform_double() [4/6]

double curand_uniform_double ( curandStateSobol64_t * state)

Return a uniformly distributed double from a Sobol64 generator.

Return a uniformly distributed double between 0.0 and 1.0 from the Sobol64 generator in state, increment position of generator. Output range excludes 0.0 but includes 1.0. Denormalized floating point outputs are never returned.

The implementation is guaranteed to use a single call to curand() to preserve the quasirandom properties of the sequence.

Parameters
state- Pointer to state to update
Returns
uniformly distributed double between 0.0 and 1.0

◆ curand_uniform_double() [5/6]

double curand_uniform_double ( curandStateTest_t * state)

◆ curand_uniform_double() [6/6]

double curand_uniform_double ( curandStateXORWOW_t * state)

Return a uniformly distributed double from an XORWOW generator.

Return a uniformly distributed double between 0.0 and 1.0 from the XORWOW generator in state, increment position of generator. Output range excludes 0.0 but includes 1.0. Denormalized floating point outputs are never returned.

The implementation may use any number of calls to curand() to get enough random bits to create the return value. The current implementation uses exactly two calls.

Parameters
state- Pointer to state to update
Returns
uniformly distributed double between 0.0 and 1.0

◆ skipahead() [1/4]

template<typename T>
void skipahead ( unsigned int n,
T state )

Update Sobol32 state to skip n elements.

Update the Sobol32 state in state to skip ahead n elements.

All values of n are valid.

Parameters
n- Number of elements to skip
state- Pointer to state to update

◆ skipahead() [2/4]

void skipahead ( unsigned long long n,
curandStateTest_t * state )

◆ skipahead() [3/4]

void skipahead ( unsigned long long n,
curandStateXORWOW_t * state )

Update XORWOW state to skip n elements.

Update the XORWOW state in state to skip ahead n elements.

All values of n are valid. Large values require more computation and so will take more time to complete.

Parameters
n- Number of elements to skip
state- Pointer to state to update

◆ skipahead() [4/4]

template<typename T>
void skipahead ( unsigned long long n,
T state )

Update Sobol64 state to skip n elements.

Update the Sobol64 state in state to skip ahead n elements.

All values of n are valid.

Parameters
n- Number of elements to skip
state- Pointer to state to update

◆ skipahead_sequence()

void skipahead_sequence ( unsigned long long n,
curandStateXORWOW_t * state )

Update XORWOW state to skip ahead n subsequences.

Update the XORWOW state in state to skip ahead n subsequences. Each subsequence is $ 2^{67} $ elements long, so this means the function will skip ahead $ 2^{67} \cdot n$ elements.

All values of n are valid. Large values require more computation and so will take more time to complete.

Parameters
n- Number of subsequences to skip
state- Pointer to state to update

Variable Documentation

◆ precalc_xorwow_matrix

unsigned int precalc_xorwow_matrix[8][2] = {{850664906UL, 4258393217UL}}
static