rocRAND examples

rocRAND examples#

rocRAND generates pseudo-random and quasi-random numbers directly into device memory on AMD GPUs. hipFORT exposes it through the hipfort_rocrand module, together with hipfort_rocrand_enums for the generator enumerations (ROCRAND_RNG_PSEUDO_PHILOX4_32_10 and so on). The programs below import both explicitly.

Every program on this page is a complete, self-contained example that is built and run as part of the hipFORT test suite. The Fortran 2008 version of each program lives in test/f2008/rocrand, and the equivalent Fortran 2003 version, which uses type(c_ptr) device pointers and explicit byte counts instead of Fortran array pointers, lives in test/f2003/rocrand.

hipRAND offers the same functionality through an API that follows cuRAND; see the hipRAND examples.

The page is organized by generator, because that is the one choice a program makes up front. Each section shows one worked example and names the sibling programs that pair the same generator with the other distributions.

rocRAND call pattern#

A rocRAND program always follows the same sequence:

  1. Create a generator with rocrand_create_generator, passing the enumerator that selects the algorithm.

  2. Configure it: pseudo-random generators take a seed via rocrand_set_seed, while quasi-random generators take a dimension count via rocrand_set_quasi_random_generator_dimensions.

  3. Allocate a device buffer to receive the numbers.

  4. Call the rocrand_generate_* routine for the distribution you want.

  5. Call hipDeviceSynchronize before reading the results.

  6. Copy the buffer back to the host.

  7. Free the device memory and release the generator with rocrand_destroy_generator.

Keep the following conventions in mind:

  • The programs generate directly into device memory. The C binding takes type(c_ptr); the programs pass c_loc(dx(1)) to it explicitly, although an array overload that accepts dx directly also exists. (rocRAND also provides host generators via rocrand_create_generator_host, which are not covered here.)

  • The count argument is the number of elements, typed integer(c_size_t), not a byte count.

  • A fixed seed makes a pseudo-random sequence reproducible, which is what lets these programs assert on sample statistics.

  • Every rocRAND call returns a status code. The examples wrap them in rocrandCheck from the hipfort_check module, which aborts on failure.

Building an example#

The examples only need the rocrand and hip hipFORT components:

find_package(hipfort REQUIRED COMPONENTS hip rocrand)

add_executable(my_rand philox_uniform.f08)
target_link_libraries(my_rand PRIVATE hipfort::rocrand hipfort::hip)

See Using hipFORT in your application for the full set of build options.

Output distributions#

Every generator on this page supports the same four combinations, which differ only in the routine called and the type of the output buffer:

  • rocrand_generate_uniform fills the buffer with real(c_float) values distributed uniformly over (0, 1], so a correct sample has a mean near 0.5. rocrand_generate_uniform_double is the real(c_double) form.

  • rocrand_generate_normal takes a mean and a standard deviation in addition to the buffer and count, and rocrand_generate_normal_double is its double-precision form.

Two further distributions, Poisson and log-normal, are exercised only with the Philox generator and are covered in their own section at the end of the page.

Pseudo-random generators#

The four pseudo-random generators are interchangeable in these programs: each is created with its own enumerator, seeded with rocrand_set_seed, and then used identically.

XORWOW#

ROCRAND_RNG_PSEUDO_XORWOW selects the xorshift-based XORWOW generator. This example generates uniform single-precision values and checks that they lie in (0, 1] and that the sample mean and standard deviation are close to 0.5 and 1/sqrt(12).

program rocrand_xorwow_uniform_test

    use iso_c_binding
    use hipfort
    use hipfort_check
    use hipfort_rocrand
    use hipfort_rocrand_enums

    implicit none

    integer(c_size_t), parameter :: N = 65536
    integer(c_int64_t), parameter :: seed = 12345_c_int64_t
    real(c_float), parameter :: expected_mean = 0.5, delta = 0.01
    real(c_float), parameter :: expected_std = 0.28867513 ! 1/sqrt(12)

    type(c_ptr) :: gen = c_null_ptr

    real(c_float), allocatable, target, dimension(:) :: hx
    real(c_float), pointer, dimension(:) :: dx
    real(c_float) :: sample_mean, sample_std

    write(*,"(a)",advance="no") "-- Running test 'XORWOW_UNIFORM' (Fortran 2008 interfaces) - "

    ! Create generator and set a fixed seed for reproducibility
    call rocrandCheck(rocrand_create_generator(gen, ROCRAND_RNG_PSEUDO_XORWOW))
    call rocrandCheck(rocrand_set_seed(gen, seed))

    ! Allocate host- and device-side memory
    allocate(hx(N))
    call hipCheck(hipMalloc(dx, mold=hx))

    ! Generate uniformly distributed floats on the device.
    ! output_data is type(c_ptr); take the address of the device pointer target.
    call rocrandCheck(rocrand_generate_uniform(gen, c_loc(dx(1)), N))
    call hipCheck(hipDeviceSynchronize())

    ! Transfer data back to host memory
    call hipCheck(hipMemcpy(hx, dx, hipMemcpyDeviceToHost))

    ! Verification: output in (0,1], with mean 1/2 and stddev 1/sqrt(12)
    if (any(hx <= 0) .or. any(hx > 1)) then
        write(*,*) "FAILED! value outside (0,1]: ", minval(hx), maxval(hx)
        call exit(1)
    end if
    sample_mean = sum(hx) / real(N)
    if (abs(sample_mean - expected_mean) > delta) then
        write(*,*) "FAILED! mean out of tolerance: ", sample_mean
        call exit(1)
    end if
    sample_std = sqrt(sum((hx - expected_mean)**2) / real(N))
    if (abs(sample_std - expected_std) > delta) then
        write(*,*) "FAILED! stddev out of tolerance: ", sample_std
        call exit(1)
    end if

    ! Cleanup
    call hipCheck(hipFree(dx))
    deallocate(hx)
    call rocrandCheck(rocrand_destroy_generator(gen))

    write(*,*) "PASSED!"

end program rocrand_xorwow_uniform_test

test/f2008/rocrand/xorwow_uniform_double.f08, xorwow_normal.f08 and xorwow_normal_double.f08 cover the remaining distributions.

Philox#

ROCRAND_RNG_PSEUDO_PHILOX4_32_10 selects the counter-based Philox generator. This example uses the normal distribution, so it also shows the extra mean and stddev arguments; it checks both the sample mean and the sample standard deviation, since a generator producing the right mean with the wrong spread would otherwise pass.

program rocrand_philox_normal_test

    use iso_c_binding
    use hipfort
    use hipfort_check
    use hipfort_rocrand
    use hipfort_rocrand_enums

    implicit none

    integer(c_size_t), parameter :: N = 65536
    integer(c_int64_t), parameter :: seed = 12345_c_int64_t
    ! Not the standard normal, so a binding that drops mean or stddev cannot pass.
    real(c_float), parameter :: mean = 2.0, stddev = 0.5, delta = 0.05

    type(c_ptr) :: gen = c_null_ptr

    real(c_float), allocatable, target, dimension(:) :: hx
    real(c_float), pointer, dimension(:) :: dx
    real(c_float) :: sample_mean, sample_std

    write(*,"(a)",advance="no") "-- Running test 'PHILOX_NORMAL' (Fortran 2008 interfaces) - "

    ! Create generator and set a fixed seed for reproducibility
    call rocrandCheck(rocrand_create_generator(gen, ROCRAND_RNG_PSEUDO_PHILOX4_32_10))
    call rocrandCheck(rocrand_set_seed(gen, seed))

    ! Allocate host- and device-side memory
    allocate(hx(N))
    call hipCheck(hipMalloc(dx, mold=hx))

    ! Generate normally distributed floats on the device.
    ! output_data is type(c_ptr); take the address of the device pointer target.
    call rocrandCheck(rocrand_generate_normal(gen, c_loc(dx(1)), N, mean, stddev))
    call hipCheck(hipDeviceSynchronize())

    ! Transfer data back to host memory
    call hipCheck(hipMemcpy(hx, dx, hipMemcpyDeviceToHost))

    ! Verification: sample mean and stddev should match requested parameters
    sample_mean = sum(hx) / real(N)
    sample_std = sqrt(sum((hx - mean)**2) / real(N))
    if (abs(sample_mean - mean) > delta) then
        write(*,*) "FAILED! mean out of tolerance: ", sample_mean
        call exit(1)
    end if
    if (abs(sample_std - stddev) > delta) then
        write(*,*) "FAILED! stddev out of tolerance: ", sample_std
        call exit(1)
    end if

    ! Cleanup
    call hipCheck(hipFree(dx))
    deallocate(hx)
    call rocrandCheck(rocrand_destroy_generator(gen))

    write(*,*) "PASSED!"

end program rocrand_philox_normal_test

test/f2008/rocrand/philox_uniform.f08, philox_uniform_double.f08 and philox_normal_double.f08 cover the remaining distributions. Philox is also the generator used for the Poisson and log-normal programs below.

MRG32K3A#

ROCRAND_RNG_PSEUDO_MRG32K3A selects the combined multiple-recursive generator.

program rocrand_mrg32k3a_uniform_test

    use iso_c_binding
    use hipfort
    use hipfort_check
    use hipfort_rocrand
    use hipfort_rocrand_enums

    implicit none

    integer(c_size_t), parameter :: N = 65536
    integer(c_int64_t), parameter :: seed = 12345_c_int64_t
    real(c_float), parameter :: expected_mean = 0.5, delta = 0.01
    real(c_float), parameter :: expected_std = 0.28867513 ! 1/sqrt(12)

    type(c_ptr) :: gen = c_null_ptr

    real(c_float), allocatable, target, dimension(:) :: hx
    real(c_float), pointer, dimension(:) :: dx
    real(c_float) :: sample_mean, sample_std

    write(*,"(a)",advance="no") "-- Running test 'MRG32K3A_UNIFORM' (Fortran 2008 interfaces) - "

    call rocrandCheck(rocrand_create_generator(gen, ROCRAND_RNG_PSEUDO_MRG32K3A))
    call rocrandCheck(rocrand_set_seed(gen, seed))

    allocate(hx(N))
    call hipCheck(hipMalloc(dx, mold=hx))

    ! Uniform (0,1] samples have mean 1/2 and stddev 1/sqrt(12).
    call rocrandCheck(rocrand_generate_uniform(gen, c_loc(dx(1)), N))
    call hipCheck(hipDeviceSynchronize())
    call hipCheck(hipMemcpy(hx, dx, hipMemcpyDeviceToHost))

    if (any(hx <= 0) .or. any(hx > 1)) then
        write(*,*) "FAILED! value outside (0,1]: ", minval(hx), maxval(hx)
        call exit(1)
    end if
    sample_mean = sum(hx) / real(N, kind(sample_mean))
    if (abs(sample_mean - expected_mean) > delta) then
        write(*,*) "FAILED! mean out of tolerance: ", sample_mean
        call exit(1)
    end if
    sample_std = sqrt(sum((hx - expected_mean)**2) / real(N, kind(sample_std)))
    if (abs(sample_std - expected_std) > delta) then
        write(*,*) "FAILED! stddev out of tolerance: ", sample_std
        call exit(1)
    end if

    call hipCheck(hipFree(dx))
    deallocate(hx)
    call rocrandCheck(rocrand_destroy_generator(gen))

    write(*,*) "PASSED!"

end program rocrand_mrg32k3a_uniform_test

test/f2008/rocrand/mrg32k3a_uniform_double.f08, mrg32k3a_normal.f08 and mrg32k3a_normal_double.f08 cover the remaining distributions.

MTGP32#

ROCRAND_RNG_PSEUDO_MTGP32 selects the Mersenne Twister for graphics processors. This example uses the normal distribution.

program rocrand_mtgp32_normal_test

    use iso_c_binding
    use hipfort
    use hipfort_check
    use hipfort_rocrand
    use hipfort_rocrand_enums

    implicit none

    integer(c_size_t), parameter :: N = 65536
    integer(c_int64_t), parameter :: seed = 12345_c_int64_t
    ! Not the standard normal, so a binding that drops mean or stddev cannot pass.
    real(c_float), parameter :: mean = 2.0, stddev = 0.5, delta = 0.05

    type(c_ptr) :: gen = c_null_ptr

    real(c_float), allocatable, target, dimension(:) :: hx
    real(c_float), pointer, dimension(:) :: dx
    real(c_float) :: sample_mean
    real(c_float) :: sample_std

    write(*,"(a)",advance="no") "-- Running test 'MTGP32_NORMAL' (Fortran 2008 interfaces) - "

    call rocrandCheck(rocrand_create_generator(gen, ROCRAND_RNG_PSEUDO_MTGP32))
    call rocrandCheck(rocrand_set_seed(gen, seed))

    allocate(hx(N))
    call hipCheck(hipMalloc(dx, mold=hx))

    ! Normal(mean, stddev) samples recover both moments.
    call rocrandCheck(rocrand_generate_normal(gen, c_loc(dx(1)), N, mean, stddev))
    call hipCheck(hipDeviceSynchronize())
    call hipCheck(hipMemcpy(hx, dx, hipMemcpyDeviceToHost))

    sample_mean = sum(hx) / real(N, kind(sample_mean))
    sample_std = sqrt(sum((hx - sample_mean)**2) / real(N, kind(sample_mean)))
    if (abs(sample_mean - mean) > delta) then
        write(*,*) "FAILED! mean out of tolerance: ", sample_mean
        call exit(1)
    end if
    if (abs(sample_std - stddev) > delta) then
        write(*,*) "FAILED! stddev out of tolerance: ", sample_std
        call exit(1)
    end if

    call hipCheck(hipFree(dx))
    deallocate(hx)
    call rocrandCheck(rocrand_destroy_generator(gen))

    write(*,*) "PASSED!"

end program rocrand_mtgp32_normal_test

test/f2008/rocrand/mtgp32_uniform.f08, mtgp32_uniform_double.f08 and mtgp32_normal_double.f08 cover the remaining distributions.

Quasi-random generators#

Sobol32#

ROCRAND_RNG_QUASI_SOBOL32 produces a low-discrepancy sequence, which fills the sample space more evenly than a pseudo-random stream. It is configured with a dimension count instead of a seed, so rocrand_set_quasi_random_generator_dimensions replaces the rocrand_set_seed call; everything else is unchanged.

program rocrand_sobol32_uniform_test

    use iso_c_binding
    use hipfort
    use hipfort_check
    use hipfort_rocrand
    use hipfort_rocrand_enums

    implicit none

    integer(c_size_t), parameter :: N = 65536
    integer(c_int), parameter :: dimensions = 1
    real(c_float), parameter :: expected_mean = 0.5, delta = 0.01
    real(c_float), parameter :: expected_std = 0.28867513 ! 1/sqrt(12)

    type(c_ptr) :: gen = c_null_ptr

    real(c_float), allocatable, target, dimension(:) :: hx
    real(c_float), pointer, dimension(:) :: dx
    real(c_float) :: sample_mean, sample_std

    write(*,"(a)",advance="no") "-- Running test 'SOBOL32_UNIFORM' (Fortran 2008 interfaces) - "

    ! Sobol32 is a quasi-random (low-discrepancy) generator: it takes a
    ! dimension count rather than a seed.
    call rocrandCheck(rocrand_create_generator(gen, ROCRAND_RNG_QUASI_SOBOL32))
    call rocrandCheck(rocrand_set_quasi_random_generator_dimensions(gen, dimensions))

    allocate(hx(N))
    call hipCheck(hipMalloc(dx, mold=hx))

    ! Uniform (0,1] samples have mean 1/2 and stddev 1/sqrt(12).
    call rocrandCheck(rocrand_generate_uniform(gen, c_loc(dx(1)), N))
    call hipCheck(hipDeviceSynchronize())
    call hipCheck(hipMemcpy(hx, dx, hipMemcpyDeviceToHost))

    if (any(hx <= 0) .or. any(hx > 1)) then
        write(*,*) "FAILED! value outside (0,1]: ", minval(hx), maxval(hx)
        call exit(1)
    end if
    sample_mean = sum(hx) / real(N, kind(sample_mean))
    if (abs(sample_mean - expected_mean) > delta) then
        write(*,*) "FAILED! mean out of tolerance: ", sample_mean
        call exit(1)
    end if
    sample_std = sqrt(sum((hx - expected_mean)**2) / real(N, kind(sample_std)))
    if (abs(sample_std - expected_std) > delta) then
        write(*,*) "FAILED! stddev out of tolerance: ", sample_std
        call exit(1)
    end if

    call hipCheck(hipFree(dx))
    deallocate(hx)
    call rocrandCheck(rocrand_destroy_generator(gen))

    write(*,*) "PASSED!"

end program rocrand_sobol32_uniform_test

test/f2008/rocrand/sobol32_uniform_double.f08, sobol32_normal.f08 and sobol32_normal_double.f08 cover the remaining distributions.

Poisson and log-normal distributions#

These two distributions are exercised with the Philox generator only, so that a failure points at the distribution entry point rather than at the generator.

Poisson#

rocrand_generate_poisson draws from a Poisson distribution with parameter lambda. Unlike the other distributions it produces unsigned 32-bit integers rather than floating-point values, so the host and device buffers are integer(c_int) while lambda remains real(c_double). A Poisson distribution has both mean and variance lambda, which is what the program checks.

program rocrand_philox_poisson_test

    use iso_c_binding
    use hipfort
    use hipfort_check
    use hipfort_rocrand
    use hipfort_rocrand_enums

    implicit none

    integer(c_size_t), parameter :: N = 65536
    integer(c_int64_t), parameter :: seed = 12345_c_int64_t
    real(c_double), parameter :: lambda = 10.0_c_double
    real(c_double), parameter :: delta = 0.1_c_double
    ! Poisson variance is lambda; allow 5% of lambda, but at least 0.5.
    real(c_double), parameter :: var_delta = max(0.05_c_double * lambda, 0.5_c_double)

    type(c_ptr) :: gen = c_null_ptr

    integer(c_int), allocatable, target, dimension(:) :: hx
    integer(c_int), pointer, dimension(:) :: dx
    real(c_double) :: sample_mean, sample_var

    write(*,"(a)",advance="no") "-- Running test 'PHILOX_POISSON' (Fortran 2008 interfaces) - "

    call rocrandCheck(rocrand_create_generator(gen, ROCRAND_RNG_PSEUDO_PHILOX4_32_10))
    call rocrandCheck(rocrand_set_seed(gen, seed))

    allocate(hx(N))
    call hipCheck(hipMalloc(dx, mold=hx))

    ! Poisson(lambda) samples are unsigned 32-bit integers with mean and variance lambda.
    call rocrandCheck(rocrand_generate_poisson(gen, c_loc(dx(1)), N, lambda))
    call hipCheck(hipDeviceSynchronize())
    call hipCheck(hipMemcpy(hx, dx, hipMemcpyDeviceToHost))

    sample_mean = real(sum(hx), c_double) / real(N, c_double)
    if (abs(sample_mean - lambda) > delta) then
        write(*,*) "FAILED! mean out of tolerance: ", sample_mean
        call exit(1)
    end if
    sample_var = sum((real(hx, c_double) - sample_mean)**2) / real(N, c_double)
    if (abs(sample_var - lambda) > var_delta) then
        write(*,*) "FAILED! variance out of tolerance: ", sample_var
        call exit(1)
    end if

    call hipCheck(hipFree(dx))
    deallocate(hx)
    call rocrandCheck(rocrand_destroy_generator(gen))

    write(*,*) "PASSED!"

end program rocrand_philox_poisson_test

This is the only rocRAND Poisson program; its Fortran 2003 version is test/f2003/rocrand/philox_poisson.f03.

Log-normal#

rocrand_generate_log_normal produces values whose logarithm is normally distributed with the given mean and standard deviation. Every sample is therefore strictly positive, and the program verifies the distribution by taking logs and checking the resulting mean and standard deviation.

program rocrand_philox_log_normal_test

    use iso_c_binding
    use hipfort
    use hipfort_check
    use hipfort_rocrand
    use hipfort_rocrand_enums

    implicit none

    integer(c_size_t), parameter :: N = 65536
    integer(c_int64_t), parameter :: seed = 12345_c_int64_t
    ! Not the standard normal, so a binding that drops mean or stddev cannot pass.
    real(c_float), parameter :: mean = 2.0, stddev = 0.5, delta = 0.05

    type(c_ptr) :: gen = c_null_ptr

    real(c_float), allocatable, target, dimension(:) :: hx
    real(c_float), pointer, dimension(:) :: dx
    real(c_float) :: sample_mean
    real(c_float) :: sample_std

    write(*,"(a)",advance="no") "-- Running test 'PHILOX_LOG_NORMAL' (Fortran 2008 interfaces) - "

    call rocrandCheck(rocrand_create_generator(gen, ROCRAND_RNG_PSEUDO_PHILOX4_32_10))
    call rocrandCheck(rocrand_set_seed(gen, seed))

    allocate(hx(N))
    call hipCheck(hipMalloc(dx, mold=hx))

    ! If x is log-normal with parameters (mean, stddev) then log(x) is
    ! normal with those moments. Log-normal samples are strictly
    ! positive, so the logarithm is always defined.
    call rocrandCheck(rocrand_generate_log_normal(gen, c_loc(dx(1)), N, mean, stddev))
    call hipCheck(hipDeviceSynchronize())
    call hipCheck(hipMemcpy(hx, dx, hipMemcpyDeviceToHost))

    if (any(hx <= 0.0)) then
        write(*,*) "FAILED! log-normal sample was not positive"
        call exit(1)
    end if
    sample_mean = sum(log(hx)) / real(N, kind(sample_mean))
    sample_std = sqrt(sum((log(hx) - sample_mean)**2) / real(N, kind(sample_mean)))
    if (abs(sample_mean - mean) > delta) then
        write(*,*) "FAILED! mean of log(x) out of tolerance: ", sample_mean
        call exit(1)
    end if
    if (abs(sample_std - stddev) > delta) then
        write(*,*) "FAILED! stddev of log(x) out of tolerance: ", sample_std
        call exit(1)
    end if

    call hipCheck(hipFree(dx))
    deallocate(hx)
    call rocrandCheck(rocrand_destroy_generator(gen))

    write(*,*) "PASSED!"

end program rocrand_philox_log_normal_test

rocrand_generate_log_normal_double is the double-precision form; see test/f2008/rocrand/philox_log_normal_double.f08.

Querying the rocRAND version#

rocrand_get_version writes the library version to a plain integer(c_int) argument, so the variable is passed directly rather than through c_loc. The value is encoded as major*100000 + minor*100 + patch.

program rocrand_version_test

    use iso_c_binding
    use hipfort_check
    use hipfort_rocrand
    use hipfort_rocrand_enums

    implicit none

    integer(c_int), parameter :: poison = -12345

    integer(c_int) :: version, version2
    integer(c_int) :: major, minor, patch

    write(*,"(a)",advance="no") "-- Running test 'rocRAND version' &
                                &(Fortran 2008 interfaces) - "

    version = poison
    version2 = poison

    call rocrandCheck(rocrand_get_version(version))
    call rocrandCheck(rocrand_get_version(version2))

    if (version == poison) then
       write(*,*) "FAILED! rocrand_get_version did not write the output argument"
       STOP 1
    end if

    if (version <= 0) then
       write(*,*) "FAILED! rocrand_get_version returned a non-positive code: ", version
       STOP 1
    end if

    major = version / 100000
    minor = mod(version / 100, 1000)
    patch = mod(version, 100)

    if (major < 1) then
       write(*,*) "FAILED! implausible rocRAND major version: ", major, " (code ", version, ")"
       STOP 1
    end if

    if (version2 /= version) then
       write(*,*) "FAILED! rocrand_get_version is not stable: ", version, " then ", version2
       STOP 1
    end if

    write(*,"(a,i0,a,i0,a,i0,a,i0,a)") " PASSED! rocRAND version: ", &
         major, ".", minor, ".", patch, " (code ", version, ")"

end program rocrand_version_test
Contents