Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Jump to bottom

[RNG] Add mcg59 #152

Open
wants to merge 7 commits into
base: develop
Choose a base branch
from
Open

[RNG] Add mcg59 #152

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
7 commits
Select commit Hold shift + click to select a range
722c312
[RNG] Add mcg59
kcenia4010 Dec 28, 2021
7382d38
[RNG] comment fix
kcenia4010 Jan 12, 2022
28f8bde
[RNG] Clang-format
kcenia4010 Jan 12, 2022
a04b09d
[RNG] code style fix
kcenia4010 Jan 20, 2022
f50b2c2
[RNG] clang format
kcenia4010 Jan 20, 2022
cd6a555
[RNG] change accessor
kcenia4010 Jan 20, 2022
a8861e7
[RNG] delete namespace
kcenia4010 Jan 20, 2022
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Jump to
Jump to file
Failed to load files.
Diff view
Diff view
16 changes: 8 additions & 8 deletions include/oneapi/mkl/rng/detail/curand/onemkl_rng_curand.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -65,10 +65,7 @@
#include "oneapi/mkl/detail/export.hpp"
#include "oneapi/mkl/rng/detail/engine_impl.hpp"

namespace oneapi {
namespace mkl {
namespace rng {
namespace curand {
namespace oneapi::mkl::rng::curand {

ONEMKL_EXPORT oneapi::mkl::rng::detail::engine_impl* create_philox4x32x10(cl::sycl::queue queue,
std::uint64_t seed);
Expand All @@ -82,9 +79,12 @@ ONEMKL_EXPORT oneapi::mkl::rng::detail::engine_impl* create_mrg32k3a(cl::sycl::q
ONEMKL_EXPORT oneapi::mkl::rng::detail::engine_impl* create_mrg32k3a(
cl::sycl::queue queue, std::initializer_list<std::uint32_t> seed);

} // namespace curand
} // namespace rng
} // namespace mkl
} // namespace oneapi
ONEMKL_EXPORT oneapi::mkl::rng::detail::engine_impl* create_mcg59(cl::sycl::queue queue,
Copy link
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

cl::sycl:: is for SYCL 1.2.1. For SYCL 2020 we can use just sycl::

Copy link
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

I think, as described in issue #149, using the sycl:: namespace is only SYCL conformant when the sycl/sycl.hpp header is included.

Copy link
Author

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

There is a PR from ROCK RAND. Will change after them merge

std::uint64_t seed);

ONEMKL_EXPORT oneapi::mkl::rng::detail::engine_impl* create_mcg59(
cl::sycl::queue queue, std::initializer_list<std::uint64_t> seed);

} // namespace oneapi::mkl::rng::curand

#endif //_ONEMKL_RNG_CURAND_HPP_
46 changes: 20 additions & 26 deletions include/oneapi/mkl/rng/detail/engine_impl.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -28,10 +28,7 @@

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

namespace oneapi {
namespace mkl {
namespace rng {
namespace detail {
namespace oneapi::mkl::rng::detail {

class engine_impl {
public:
Expand All @@ -41,58 +38,58 @@ class engine_impl {

// Buffers API
virtual void generate(const uniform<float, uniform_method::standard>& distr, std::int64_t n,
cl::sycl::buffer<float, 1>& r) = 0;
cl::sycl::buffer<float>& r) = 0;

virtual void generate(const uniform<double, uniform_method::standard>& distr, std::int64_t n,
cl::sycl::buffer<double, 1>& r) = 0;
cl::sycl::buffer<double>& r) = 0;

virtual void generate(const uniform<std::int32_t, uniform_method::standard>& distr,
std::int64_t n, cl::sycl::buffer<std::int32_t, 1>& r) = 0;
std::int64_t n, cl::sycl::buffer<std::int32_t>& r) = 0;

virtual void generate(const uniform<float, uniform_method::accurate>& distr, std::int64_t n,
cl::sycl::buffer<float, 1>& r) = 0;
cl::sycl::buffer<float>& r) = 0;

virtual void generate(const uniform<double, uniform_method::accurate>& distr, std::int64_t n,
cl::sycl::buffer<double, 1>& r) = 0;
cl::sycl::buffer<double>& r) = 0;

virtual void generate(const gaussian<float, gaussian_method::box_muller2>& distr,
std::int64_t n, cl::sycl::buffer<float, 1>& r) = 0;
std::int64_t n, cl::sycl::buffer<float>& r) = 0;

virtual void generate(const gaussian<double, gaussian_method::box_muller2>& distr,
std::int64_t n, cl::sycl::buffer<double, 1>& r) = 0;
std::int64_t n, cl::sycl::buffer<double>& r) = 0;

virtual void generate(const gaussian<float, gaussian_method::icdf>& distr, std::int64_t n,
cl::sycl::buffer<float, 1>& r) = 0;
cl::sycl::buffer<float>& r) = 0;

virtual void generate(const gaussian<double, gaussian_method::icdf>& distr, std::int64_t n,
cl::sycl::buffer<double, 1>& r) = 0;
cl::sycl::buffer<double>& r) = 0;

virtual void generate(const lognormal<float, lognormal_method::box_muller2>& distr,
std::int64_t n, cl::sycl::buffer<float, 1>& r) = 0;
std::int64_t n, cl::sycl::buffer<float>& r) = 0;

virtual void generate(const lognormal<double, lognormal_method::box_muller2>& distr,
std::int64_t n, cl::sycl::buffer<double, 1>& r) = 0;
std::int64_t n, cl::sycl::buffer<double>& r) = 0;

virtual void generate(const lognormal<float, lognormal_method::icdf>& distr, std::int64_t n,
cl::sycl::buffer<float, 1>& r) = 0;
cl::sycl::buffer<float>& r) = 0;

virtual void generate(const lognormal<double, lognormal_method::icdf>& distr, std::int64_t n,
cl::sycl::buffer<double, 1>& r) = 0;
cl::sycl::buffer<double>& r) = 0;

virtual void generate(const bernoulli<std::int32_t, bernoulli_method::icdf>& distr,
std::int64_t n, cl::sycl::buffer<std::int32_t, 1>& r) = 0;
std::int64_t n, cl::sycl::buffer<std::int32_t>& r) = 0;

virtual void generate(const bernoulli<std::uint32_t, bernoulli_method::icdf>& distr,
std::int64_t n, cl::sycl::buffer<std::uint32_t, 1>& r) = 0;
std::int64_t n, cl::sycl::buffer<std::uint32_t>& r) = 0;

virtual void generate(const poisson<std::int32_t, poisson_method::gaussian_icdf_based>& distr,
std::int64_t n, cl::sycl::buffer<std::int32_t, 1>& r) = 0;
std::int64_t n, cl::sycl::buffer<std::int32_t>& r) = 0;

virtual void generate(const poisson<std::uint32_t, poisson_method::gaussian_icdf_based>& distr,
std::int64_t n, cl::sycl::buffer<std::uint32_t, 1>& r) = 0;
std::int64_t n, cl::sycl::buffer<std::uint32_t>& r) = 0;

virtual void generate(const bits<std::uint32_t>& distr, std::int64_t n,
cl::sycl::buffer<std::uint32_t, 1>& r) = 0;
cl::sycl::buffer<std::uint32_t>& r) = 0;

// USM APIs
virtual cl::sycl::event generate(const uniform<float, uniform_method::standard>& distr,
Expand Down Expand Up @@ -185,9 +182,6 @@ class engine_impl {
cl::sycl::queue queue_;
};

} // namespace detail
} // namespace rng
} // namespace mkl
} // namespace oneapi
} // namespace oneapi::mkl::rng::detail

#endif //_ONEMKL_RNG_ENGINE_IMPL_HPP_
16 changes: 8 additions & 8 deletions include/oneapi/mkl/rng/detail/mklcpu/onemkl_rng_mklcpu.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -26,10 +26,7 @@
#include "oneapi/mkl/detail/export.hpp"
#include "oneapi/mkl/rng/detail/engine_impl.hpp"

namespace oneapi {
namespace mkl {
namespace rng {
namespace mklcpu {
namespace oneapi::mkl::rng::mklcpu {

ONEMKL_EXPORT oneapi::mkl::rng::detail::engine_impl* create_philox4x32x10(cl::sycl::queue queue,
std::uint64_t seed);
Expand All @@ -43,9 +40,12 @@ ONEMKL_EXPORT oneapi::mkl::rng::detail::engine_impl* create_mrg32k3a(cl::sycl::q
ONEMKL_EXPORT oneapi::mkl::rng::detail::engine_impl* create_mrg32k3a(
cl::sycl::queue queue, std::initializer_list<std::uint32_t> seed);

} // namespace mklcpu
} // namespace rng
} // namespace mkl
} // namespace oneapi
ONEMKL_EXPORT oneapi::mkl::rng::detail::engine_impl* create_mcg59(cl::sycl::queue queue,
std::uint64_t seed);

ONEMKL_EXPORT oneapi::mkl::rng::detail::engine_impl* create_mcg59(
cl::sycl::queue queue, std::initializer_list<std::uint64_t> seed);

} // namespace oneapi::mkl::rng::mklcpu

#endif //_ONEMKL_RNG_MKLCPU_HPP_
16 changes: 8 additions & 8 deletions include/oneapi/mkl/rng/detail/mklgpu/onemkl_rng_mklgpu.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -26,10 +26,7 @@
#include "oneapi/mkl/detail/export.hpp"
#include "oneapi/mkl/rng/detail/engine_impl.hpp"

namespace oneapi {
namespace mkl {
namespace rng {
namespace mklgpu {
namespace oneapi::mkl::rng::mklgpu {

ONEMKL_EXPORT oneapi::mkl::rng::detail::engine_impl* create_philox4x32x10(cl::sycl::queue queue,
std::uint64_t seed);
Expand All @@ -43,9 +40,12 @@ ONEMKL_EXPORT oneapi::mkl::rng::detail::engine_impl* create_mrg32k3a(cl::sycl::q
ONEMKL_EXPORT oneapi::mkl::rng::detail::engine_impl* create_mrg32k3a(
cl::sycl::queue queue, std::initializer_list<std::uint32_t> seed);

} // namespace mklgpu
} // namespace rng
} // namespace mkl
} // namespace oneapi
ONEMKL_EXPORT oneapi::mkl::rng::detail::engine_impl* create_mcg59(cl::sycl::queue queue,
std::uint64_t seed);

ONEMKL_EXPORT oneapi::mkl::rng::detail::engine_impl* create_mcg59(
cl::sycl::queue queue, std::initializer_list<std::uint64_t> seed);

} // namespace oneapi::mkl::rng::mklgpu

#endif //_ONEMKL_RNG_MKLGPU_HPP_
16 changes: 8 additions & 8 deletions include/oneapi/mkl/rng/detail/rng_loader.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -28,10 +28,7 @@

#include "oneapi/mkl/rng/detail/engine_impl.hpp"

namespace oneapi {
namespace mkl {
namespace rng {
namespace detail {
namespace oneapi::mkl::rng::detail {

ONEMKL_EXPORT engine_impl* create_philox4x32x10(oneapi::mkl::device libkey, cl::sycl::queue queue,
std::uint64_t seed);
Expand All @@ -45,9 +42,12 @@ ONEMKL_EXPORT engine_impl* create_mrg32k3a(oneapi::mkl::device libkey, cl::sycl:
ONEMKL_EXPORT engine_impl* create_mrg32k3a(oneapi::mkl::device libkey, cl::sycl::queue queue,
std::initializer_list<std::uint32_t> seed);

} // namespace detail
} // namespace rng
} // namespace mkl
} // namespace oneapi
ONEMKL_EXPORT engine_impl* create_mcg59(oneapi::mkl::device libkey, cl::sycl::queue queue,
std::uint64_t seed);

ONEMKL_EXPORT engine_impl* create_mcg59(oneapi::mkl::device libkey, cl::sycl::queue queue,
std::initializer_list<std::uint64_t> seed);

} // namespace oneapi::mkl::rng::detail

#endif //_ONEMKL_RNG_LOADER_HPP_
12 changes: 5 additions & 7 deletions include/oneapi/mkl/rng/distributions.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -26,9 +26,7 @@

#include "oneapi/mkl/exceptions.hpp"

namespace oneapi {
namespace mkl {
namespace rng {
namespace oneapi::mkl::rng {

// Class template oneapi::mkl::rng::uniform
//
Expand Down Expand Up @@ -358,12 +356,12 @@ class poisson {
template <typename UIntType = std::uint32_t>
class bits {
public:
static_assert(std::is_same<UIntType, std::uint32_t>::value, "rng bits type is not supported");
static_assert(std::is_same<UIntType, std::uint32_t>::value ||
std::is_same<UIntType, std::uint64_t>::value,
"rng bits type is not supported");
using result_type = UIntType;
};

} // namespace rng
} // namespace mkl
} // namespace oneapi
} // namespace oneapi::mkl::rng

#endif //_ONEMKL_RNG_DISTRIBUTIONS_HPP_
103 changes: 85 additions & 18 deletions include/oneapi/mkl/rng/engines.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -41,9 +41,7 @@
#include "oneapi/mkl/rng/detail/curand/onemkl_rng_curand.hpp"
#endif

namespace oneapi {
namespace mkl {
namespace rng {
namespace oneapi::mkl::rng {

// Class oneapi::mkl::rng::philox4x32x10
//
Expand All @@ -55,10 +53,10 @@ class philox4x32x10 {
public:
static constexpr std::uint64_t default_seed = 0;

philox4x32x10(sycl::queue queue, std::uint64_t seed = default_seed)
philox4x32x10(cl::sycl::queue queue, std::uint64_t seed = default_seed)
: pimpl_(detail::create_philox4x32x10(get_device_id(queue), queue, seed)) {}

philox4x32x10(sycl::queue queue, std::initializer_list<std::uint64_t> seed)
philox4x32x10(cl::sycl::queue queue, std::initializer_list<std::uint64_t> seed)
: pimpl_(detail::create_philox4x32x10(get_device_id(queue), queue, seed)) {}

#ifdef ENABLE_MKLCPU_BACKEND
Expand Down Expand Up @@ -121,12 +119,12 @@ class philox4x32x10 {

template <typename Distr, typename Engine>
friend void generate(const Distr& distr, Engine& engine, std::int64_t n,
sycl::buffer<typename Distr::result_type, 1>& r);
cl::sycl::buffer<typename Distr::result_type>& r);

template <typename Distr, typename Engine>
friend sycl::event generate(const Distr& distr, Engine& engine, std::int64_t n,
typename Distr::result_type* r,
const std::vector<sycl::event>& dependencies);
friend cl::sycl::event generate(const Distr& distr, Engine& engine, std::int64_t n,
typename Distr::result_type* r,
const std::vector<cl::sycl::event>& dependencies);
};

// Class oneapi::mkl::rng::mrg32k3a
Expand All @@ -139,10 +137,10 @@ class mrg32k3a {
public:
static constexpr std::uint32_t default_seed = 1;

mrg32k3a(sycl::queue queue, std::uint32_t seed = default_seed)
mrg32k3a(cl::sycl::queue queue, std::uint32_t seed = default_seed)
: pimpl_(detail::create_mrg32k3a(get_device_id(queue), queue, seed)) {}

mrg32k3a(sycl::queue queue, std::initializer_list<std::uint32_t> seed)
mrg32k3a(cl::sycl::queue queue, std::initializer_list<std::uint32_t> seed)
: pimpl_(detail::create_mrg32k3a(get_device_id(queue), queue, seed)) {}

#ifdef ENABLE_MKLCPU_BACKEND
Expand Down Expand Up @@ -202,19 +200,88 @@ class mrg32k3a {

template <typename Distr, typename Engine>
friend void generate(const Distr& distr, Engine& engine, std::int64_t n,
sycl::buffer<typename Distr::result_type, 1>& r);
cl::sycl::buffer<typename Distr::result_type>& r);

template <typename Distr, typename Engine>
friend sycl::event generate(const Distr& distr, Engine& engine, std::int64_t n,
typename Distr::result_type* r,
const std::vector<sycl::event>& dependencies);
friend cl::sycl::event generate(const Distr& distr, Engine& engine, std::int64_t n,
typename Distr::result_type* r,
const std::vector<cl::sycl::event>& dependencies);
};

// Class oneapi::mkl::rng::mcg59
//
// Represents MCG59 counter-based pseudorandom number generator
//
// Supported parallelization methods:
// leapfrog
class mcg59 {
public:
static constexpr std::uint64_t default_seed = 0;

mcg59(cl::sycl::queue queue, std::uint64_t seed = default_seed)
: pimpl_(detail::create_mcg59(get_device_id(queue), queue, seed)) {}

#ifdef ENABLE_MKLCPU_BACKEND
mcg59(backend_selector<backend::mklcpu> selector, std::uint64_t seed = default_seed)
: pimpl_(mklcpu::create_mcg59(selector.get_queue(), seed)) {}

#endif

#ifdef ENABLE_MKLGPU_BACKEND
mcg59(backend_selector<backend::mklgpu> selector, std::uint64_t seed = default_seed)
: pimpl_(mklgpu::create_mcg59(selector.get_queue(), seed)) {}

#endif

#ifdef ENABLE_CURAND_BACKEND
mcg59(backend_selector<backend::curand> selector, std::uint64_t seed = default_seed)
: pimpl_(curand::create_mcg59(selector.get_queue(), seed)) {}
#endif

mcg59(const mcg59& other) {
pimpl_.reset(other.pimpl_.get()->copy_state());
}

mcg59(mcg59&& other) {
pimpl_ = std::move(other.pimpl_);
}

mcg59& operator=(const mcg59& other) {
if (this == &other)
return *this;
pimpl_.reset(other.pimpl_.get()->copy_state());
return *this;
}

mcg59& operator=(mcg59&& other) {
if (this == &other)
return *this;
pimpl_ = std::move(other.pimpl_);
return *this;
}

private:
std::unique_ptr<detail::engine_impl> pimpl_;

template <typename Engine>
friend void skip_ahead(Engine& engine, std::uint64_t num_to_skip);

template <typename Engine>
friend void leapfrog(Engine& engine, std::uint64_t idx, std::uint64_t stride);

template <typename Distr, typename Engine>
friend void generate(const Distr& distr, Engine& engine, std::int64_t n,
cl::sycl::buffer<typename Distr::result_type>& r);

template <typename Distr, typename Engine>
friend cl::sycl::event generate(const Distr& distr, Engine& engine, std::int64_t n,
typename Distr::result_type* r,
const std::vector<cl::sycl::event>& dependencies);
};

// Default engine to be used for common cases
using default_engine = philox4x32x10;

} // namespace rng
} // namespace mkl
} // namespace oneapi
} // namespace oneapi::mkl::rng

#endif //_ONEMKL_RNG_ENGINES_HPP_