SYCL Kernels API

Supported C++ Standard Library APIs and Algorithms

oneDPL defines a subset of the C++ Standard library APIs for use in SYCL kernels. These APIs can be employed in the kernels similarly to how they are employed in code for a typical CPU-based platform.

Random Number Generation

oneDPL provides a subset of the standard C++ pseudo-random number generation functionality suitable to use within SYCL kernels. The APIs are defined in the <oneapi/dpl/random> header.

Supported Functionality

• Engine class templates:

  • linear_congruential_engine

  • subtract_with_carry_engine

• Engine adaptor class templates:

  • discard_block_engine

• Engines and engine adaptors with predefined parameters:

  • minstd_rand0

  • minstd_rand

  • ranlux24_base

  • ranlux48_base

  • ranlux24

  • ranlux48

• Distribution class templates:

  • uniform_int_distribution

  • uniform_real_distribution

  • normal_distribution

  • exponential_distribution

  • bernoulli_distribution

  • geometric_distribution

  • weibull_distribuion

  • lognormal_distribution

  • cauchy_distribution

  • extreme_value_distribution

linear_congruential_engine and subtract_with_carry_engine satisfy the uniform random bit generator requirements.

Limitations

The following deviations from the C++ Standard may apply:

• random_device and seed_seq classes and related APIs in other classes are not required;

• specifying the size of a random number engine’s state is not required;

• distributions are only required to operate with floating point types applicable to supported SYCL devices.

Extensions

As an extension to the C++ Standard, sycl::vec<Type, N> can be used as the data type template parameter for engines, engine adaptors, and distributions, where Type is one of data types supported by the corresponding class template in the standard. For such template instantiations, the result_type is also defined to sycl::vec<Type, N>.

Engines, engine adaptors, and distributions additionally define scalar_type, equivalent to the following:

• using scalar_type = typename result_type::element_type; if result_type is sycl::vec<Type, N>,

• otherwise, using scalar_type = result_type;

The scalar_type is used instead of result_type in all contexts where a scalar data type is expected, including

• the type of configuration parameters and properties,

• the seed value type,

• the input parameters of constructors,

• the return value type of min() and max() methods, etc.

Since scalar_type is the same as result_type except for template instantiations with sycl::vec, class templates still meet the applicable requirements of the C++ Standard.

When instantiated with sycl::vec<Type,N>, linear_congruential_engine and subtract_with_carry_engine may not formally satisfy the uniform random bit generator requirements defined by the C++ Standard. Instead, the following alternative requirements apply: for an engine object g of type G,

• G::scalar_type is an unsigned integral type same as sycl::vec<Type,N>::element_type,

• G::min() and G::max() return a value of G::scalar_type,

• for each index i in the range [0, N), G::min() <= g()[i] and g()[i] <= G::max().

Effectively, these engines satisfy the standard uniform random bit generator requirements for each element of a sycl::vec returned by their operator().

Similarly, for a distribution object d of a type D that is a template instantiated with sycl::vec<Type,N>:

• D::scalar_type is the same as sycl::vec<Type,N>::element_type,

• D::min() and D::max() return a value of D::scalar_type, and D::min() <= D::max(),

• operator() of a distribution returns a sycl::vec<Type,N> filled with random values in the closed interval [D::min(), D::max()];

The following engines and engine adaptors with predefined parameters are defined:

template <int N>
using minstd_rand0_vec = linear_congruential_engine<sycl::vec<::std::uint_fast32_t, N>, 16807, 0, 2147483647>;

template <int N>
using minstd_rand_vec = linear_congruential_engine<sycl::vec<uint_fast32_t, N>, 48271, 0, 2147483647>;

template <int N>
using ranlux24_base_vec = subtract_with_carry_engine<sycl::vec<uint_fast32_t, N>, 24, 10, 24>;

template <int N>
using ranlux48_base_vec = subtract_with_carry_engine<sycl::vec<uint_fast64_t, N>, 48, 5, 12>;

template <int N>
using ranlux24_vec = discard_block_engine<ranlux24_base_vec<N>, 223, 23>;

template <int N>
using ranlux48_vec = discard_block_engine<ranlux48_base_vec<N>, 389, 11>;

Except for producing a sycl::vec of random values per invocation, the behavior of these engines is equivalent to the corresponding scalar engines, as described in the following table:

Engines and engine adaptors based on sycl::vec<>	C++ standard analogue	The 10000th scalar random value consecutively produced by a default-constructed object
minstd_rand0_vec	minstd_rand0	1043618065
minstd_rand_vec	minstd_rand	399268537
ranlux24_base_vec	ranlux24_base	7937952
ranlux48_base_vec	ranlux48_base	61839128582725
ranlux24_vec	ranlux24	9901578
ranlux48_vec	ranlux48	1112339016

Function Objects

The oneDPL function objects are defined in the <oneapi/dpl/functional> header.

namespace oneapi {
namespace dpl {
    struct identity
    {
        template <typename T>
        constexpr T&&
        operator()(T&& t) const noexcept;
    };
}
}

The oneapi::dpl::identity class implements an identity operation. Its function operator receives an instance of a type and returns the argument unchanged.