// Copyright (c) Advanced Micro Devices, Inc., or its affiliates. // SPDX-License-Identifier: MIT #ifndef GUARD_RANDOM_GEN_ #define GUARD_RANDOM_GEN_ #include #include #include #include #include MIOPEN_DECLARE_ENV_VAR_UINT64(MIOPEN_DEBUG_DRIVER_PRNG_SEED, 12345678) namespace env = miopen::env; namespace prng { namespace details { using glibc_gen = std::linear_congruential_engine; inline std::random_device::result_type get_default_seed() { static std::random_device::result_type seed{[] { auto external_seed = env::value(MIOPEN_DEBUG_DRIVER_PRNG_SEED); auto seed_ = external_seed == 0 ? std::random_device{}() : static_cast(external_seed); // Check if JSON mode is enabled in MIOPEN_PERFORMANCE_LOGS if(miopen::IsPerformanceLoggingEnabled()) { std::cout << "{\"prng_seed\":" << seed_ << "}" << std::endl; } else { std::cout << "PRNG seed: " << seed_ << "\n"; } return seed_; }()}; return seed; } inline glibc_gen& get_prng() { static thread_local glibc_gen gen{get_default_seed()}; return gen; } template struct has_digits : std::false_type { }; template struct has_digits::digits)>> : std::true_type { }; } // namespace details inline void reset_seed(std::random_device::result_type seed = 0) { details::get_prng().seed(seed + details::get_default_seed()); } // similar to std::generate_canonical, but simpler and faster template inline T gen_canonical() { if constexpr(std::is_floating_point_v) // native fp { static constexpr T range = static_cast(1) / static_cast(details::glibc_gen::max() - details::glibc_gen::min() + 1); return range * static_cast(details::get_prng()() - details::glibc_gen::min()); } else if constexpr(std::is_integral_v) { auto val = details::get_prng()(); return static_cast(((val >> 4) + (val >> 16)) & 0x1); } else { return static_cast(gen_canonical()); } } template inline T gen_0_to_B(T B) { if constexpr(std::is_floating_point_v) // native fp { return gen_canonical() * B; } else if constexpr(std::is_integral_v) { // can only generate 27bit range, so it may not be suitable // for huge 64 bit ranges, but we do not expect such ranges return static_cast((details::get_prng()() >> 4) % B); } else // half/bfloat/etc { return static_cast(gen_0_to_B(static_cast(B))); } } template inline T gen_A_to_B(T A, T B) { assert(B > A); return gen_0_to_B(B - A) + A; } template inline T gen_off_range(T offset, T range) { static_assert(std::is_integral_v); return prng::gen_0_to_B(range) + offset; } template inline T gen_subnorm() { T denorm_val = static_cast(0); if constexpr(!std::is_integral_v && !std::is_same_v && std::is_trivially_copyable::value && details::has_digits::value) { using BitType = std::conditional_t>; static_assert(sizeof(T) == sizeof(BitType)); // -1 because ::digits counts the first implicit digit static constexpr auto mantissa_bits = std::numeric_limits::digits - 1; BitType denorm_bits = static_cast(gen_0_to_B(1 << mantissa_bits)); denorm_bits |= Signed ? (gen_canonical() << (sizeof(T) * 8 - 1)) : 0; // the proper way to do a type punning std::memcpy(&denorm_val, &denorm_bits, sizeof(T)); } return denorm_val; } } // namespace prng #endif // GUARD_RANDOM_GEN_