Intel® oneAPI Math Kernel Library (oneMKL) - Data Parallel C++ Developer Reference

Developer Reference

Contents

oneMKL RNG Device Usage Model

A typical usage model for device routines is the same as described in oneMKL RNG Usage Model:
  1. Create and initialize the object for basic random number generator.
  2. Create and initialize the object for distribution generator.
  3. Call the generate routine to get random numbers with appropriate statistical distribution.

Example of Scalar Random Numbers Generation

#include<iostream>
#include <CL/sycl.hpp>
#include "oneapi/mkl/rng/device.hpp"


int main() {
    sycl::queue queue;
    const int n = 1000;
    const int seed = 1;
    // Prepare an array for random numbers
    std::vector<float> r(n);


    sycl::buffer<float, 1> r_buf(r.data(), r.size());
    // Submit a kernel to generate on device
    queue.submit([&](sycl::handler& cgh) {
        auto r_acc = r_buf.template get_access<sycl::access::mode::write>(cgh);
        cgh.parallel_for(sycl::range<1>(n), [=](sycl::item<1> item) {
            // Create an engine object
            oneapi::mkl::rng::device::philox4x32x10<> engine(seed, item.get_id(0));
            // Create a distribution object
            oneapi::mkl::rng::device::uniform<> distr;
            // Call generate function to obtain scalar random number
            float res = oneapi::mkl::rng::device::generate(distr, engine);


            r_acc[item.get_id(0)] = res;
        });
    });


    auto r_acc = r_buf.template get_access<sycl::access::mode::read>();
    std::cout << "Samples of uniform distribution" << std::endl;
    for(int i = 0; i < 10; i++) {
        std::cout << r_acc[i] << std::endl;
    }


    return 0;
}

Example of Vector Random Numbers Generation

#include<iostream>
#include <CL/sycl.hpp>


#include "oneapi/mkl/rng/device.hpp"


int main() {
    sycl::queue queue;
    const int n = 1000;
    const int seed = 1;
    const int vec_size = 4;
    // Prepare an array for random numbers
    std::vector<float> r(n);


    sycl::buffer<float, 1> r_buf(r.data(), r.size());
    // Submit a kernel to generate on device
    sycl::queue{}.submit([&](sycl::handler& cgh) {
        auto r_acc = r_buf.template get_access<sycl::access::mode::write>(cgh);
        cgh.parallel_for(sycl::range<1>(n / vec_size), [=](sycl::item<1> item) {
            // Create an engine object
            oneapi::mkl::rng::device::philox4x32x10<vec_size> engine(seed, item.get_id(0) * vec_size);
            // Create a distribution object
            oneapi::mkl::rng::device::uniform<> distr;
            // Call generate function to obtain sycl::vec<float, 4> with random numbers
            auto res = oneapi::mkl::rng::device::generate(distr, engine);


            res.store(ite.get_id(0), r_acc);
        });
    });


    auto r_acc = r_buf.template get_access<sycl::access::mode::read>();
    std::cout << "Samples of uniform distribution" << std::endl;
    for(int i = 0; i < 10; i++) {
        std::cout << r_acc[i] << std::endl;
    }


    return 0;
}
There is an opportunity to store engines between kernels manually via
sycl::buffer
 / USM pointers or by using a specific host-side helper class called, engine descriptor.

Example of Random Numbers Generation by Engines Stored in
sycl::buffer

Engines are initialized in the first kernel. Random number generation is performed in the second kernel.
#include<iostream>
#include <CL/sycl.hpp>


#include "oneapi/mkl/rng/device.hpp"


int main() {
    sycl::queue queue;
    const int n = 1000;
    const int seed = 1;
    const int vec_size = 4;
    // Prepare an array for random numbers
    std::vector<float> r(n);
    sycl::buffer<float, 1> r_buf(r.data(), r.size());
    sycl::range<1> range(n / vec_size);
    sycl::buffer<oneapi::mkl::rng::device::mrg32k3a<vec_size>, 1> engine_buf(range);


    sycl::queue queue;


    // Kernel with initialization of engines
    queue.submit([&](sycl::handler& cgh) {
        // Create an accessor to sycl::buffer with engines to write initialized states
        auto engine_acc = engine_buf.template get_access<sycl::access::mode::write>(cgh);
        cgh.parallel_for(range, [=](sycl::item<1> item) {
            size_t id = item.get_id(0);
            // Create an engine object with offset id * 2^64
            oneapi::mkl::rng::device::mrg32k3a<vec_size> engine(seed, {0, id});
            engine_acc[id] = engine;
        });
    });
    // Kernel for random numbers generation
    queue.submit([&](sycl::handler& cgh) {
        auto r_acc = r_buf.template get_access<sycl::access::mode::write>(cgh);
        // Create an accessor to sycl::buffer with engines to read initialized states
        auto engine_acc = engine_buf.template get_access<sycl::access::mode::read>(cgh);
        cgh.parallel_for(range, [=](sycl::item<1> item) {
            size_t id = item.get_id(0);


            auto engine = engine_acc[id];
            oneapi::mkl::rng::device::uniform distr;
            auto res = oneapi::mkl::rng::device::generate(distr, engine);


            res.store(id, r_acc);
        });
    });


    auto r_acc = r_buf.template get_access<sycl::access::mode::read>();
    std::cout << "Samples of uniform distribution" << std::endl;
    for(int i = 0; i < 10; i++) {
        std::cout << r_acc[i] << std::endl;
    }


    return 0;
}

Example of Random Numbers Generation with Host-side Helpers Usage

#include<iostream>
#include <CL/sycl.hpp>


#include "oneapi/mkl/rng/device.hpp"


int main() {
    sycl::queue queue;
    const int n = 1000;
    const int seed = 1;
    const int vec_size = 4;
    // prepare array for random numbers
    std::vector<float> r(n);
    sycl::buffer<float, 1> r_buf(r.data(), r.size());
    sycl::range<1> range(n / vec_size);


    // offset of each engine in engine_descriptor
    int offset = vec_size;
    // each engine would be created in enqueued task as of specified range
    // as oneapi::mkl::rng::device::mrg32k3a<vec_size>(seed, id * offset);
    oneapi::mkl::rng::device::engine_descriptor<oneapi::mkl::rng::device::mrg32k3a<vec_size>>
    descr(queue, range, seed, offset);
    queue.submit([&](sycl::handler& cgh) {
        auto r_acc = r_buf.template get_access<sycl::access::mode::write>(cgh);
        // create engine_accessor
        auto engine_acc = descr.get_access(cgh);
        cgh.parallel_for(range, [=](sycl::item<1> item) {
            size_t id = item.get_id(0);
            // load engine from engine_accessor
            auto engine = engine_acc.load(id);
            oneapi::mkl::rng::device::uniform<Type> distr;


            auto res = oneapi::mkl::rng::device::generate(distr, engine);


            res.store(id, r_acc);
            // store engine for furter calculations if needed
            engine_acc.store(engine, id);
        });
    });


    auto r_acc = r_buf.template get_access<sycl::access::mode::read>();
    std::cout << "Samples of uniform distribution" << std::endl;
    for(int i = 0; i < 10; i++) {
        std::cout << r_acc[i] << std::endl;
    }


    return 0;
}
Additionally, examples that demonstrate usage of random number generators device routines are available in:
${MKL}/examples/dpcpp_device/rng/source

Product and Performance Information

1

Performance varies by use, configuration and other factors. Learn more at www.Intel.com/PerformanceIndex.