Visible to Intel only — GUID: GUID-BA9119AB-2FF2-426D-B9AE-319FA8FD46AE
Visible to Intel only — GUID: GUID-BA9119AB-2FF2-426D-B9AE-319FA8FD46AE
oneMKL RNG Device Usage Model
Example of Random Numbers Generation by Engines Stored in sycl::buffer
Example of Random Numbers Generation with Host-side Helpers Usage
A typical usage model for device routines is the same as described in oneMKL RNG Usage Model:
Create and initialize the object for basic random number generator.
Create and initialize the object for distribution generator.
Call the generate routine to get random numbers with appropriate statistical distribution.
Example of Scalar Random Numbers Generation
#include <iostream>
#include <sycl/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 <sycl/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 <sycl/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 <sycl/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