From b4403d63f85e758166ebb410240dd2dff41ef860 Mon Sep 17 00:00:00 2001 From: Matt Borland Date: Mon, 3 Aug 2026 11:34:51 -0400 Subject: [PATCH 1/5] Add byte conversion functions --- include/boost/int128.hpp | 1 + include/boost/int128/byte_conversions.hpp | 366 ++++++++++++++++++++++ 2 files changed, 367 insertions(+) create mode 100644 include/boost/int128/byte_conversions.hpp diff --git a/include/boost/int128.hpp b/include/boost/int128.hpp index 7d996bd7..80315fd3 100644 --- a/include/boost/int128.hpp +++ b/include/boost/int128.hpp @@ -7,6 +7,7 @@ #include #include +#include #include #include #include diff --git a/include/boost/int128/byte_conversions.hpp b/include/boost/int128/byte_conversions.hpp new file mode 100644 index 00000000..97abca1e --- /dev/null +++ b/include/boost/int128/byte_conversions.hpp @@ -0,0 +1,366 @@ +// Copyright 2026 Matt Borland +// Distributed under the Boost Software License, Version 1.0. +// https://www.boost.org/LICENSE_1_0.txt + +#ifndef BOOST_INT128_BYTE_CONVERSIONS_HPP +#define BOOST_INT128_BYTE_CONVERSIONS_HPP + +#include +#include +#include + +#ifndef BOOST_INT128_BUILD_MODULE + +#include +#include +#include +#include + +#endif + +namespace boost { +namespace int128 { + +//===================================== +// Whole value byte order conversions +//===================================== + +BOOST_INT128_EXPORT BOOST_INT128_HOST_DEVICE constexpr uint128 to_be(const uint128 value) noexcept +{ + #if BOOST_INT128_ENDIAN_BIG_BYTE + + return value; + + #else + + return byteswap(value); + + #endif +} + +BOOST_INT128_EXPORT BOOST_INT128_HOST_DEVICE constexpr int128 to_be(const int128 value) noexcept +{ + #if BOOST_INT128_ENDIAN_BIG_BYTE + + return value; + + #else + + // Reversing the two's complement bit pattern is the same operation for both signs + return static_cast(byteswap(static_cast(value))); + + #endif +} + +BOOST_INT128_EXPORT BOOST_INT128_HOST_DEVICE constexpr uint128 from_be(const uint128 value) noexcept +{ + // Self-inverse + return to_be(value); +} + +BOOST_INT128_EXPORT BOOST_INT128_HOST_DEVICE constexpr int128 from_be(const int128 value) noexcept +{ + // Self-inverse + return to_be(value); +} + +BOOST_INT128_EXPORT BOOST_INT128_HOST_DEVICE constexpr uint128 to_le(const uint128 value) noexcept +{ + #if BOOST_INT128_ENDIAN_LITTLE_BYTE + + return value; + + #else + + return byteswap(value); + + #endif +} + +BOOST_INT128_EXPORT BOOST_INT128_HOST_DEVICE constexpr int128 to_le(const int128 value) noexcept +{ + #if BOOST_INT128_ENDIAN_LITTLE_BYTE + + return value; + + #else + + // Reversing the two's complement bit pattern is the same operation for both signs + return static_cast(byteswap(static_cast(value))); + + #endif +} + +BOOST_INT128_EXPORT BOOST_INT128_HOST_DEVICE constexpr uint128 from_le(const uint128 value) noexcept +{ + // Self-inverse + return to_le(value); +} + +BOOST_INT128_EXPORT BOOST_INT128_HOST_DEVICE constexpr int128 from_le(const int128 value) noexcept +{ + // Self-inverse + return to_le(value); +} + +namespace detail { + +// The byte-like element types that the byte array functions operate on. +// std::byte only exists from C++17, and the library is usable from C++14. +template +struct byte_like +{ + static constexpr bool value = std::is_same::value || + std::is_same::value || + std::is_same::value + #if defined(__cpp_lib_byte) && __cpp_lib_byte >= 201603L + || std::is_same::value + #endif + ; +}; + +template +BOOST_INT128_INLINE_CONSTEXPR bool is_byte_like_v = byte_like::value; + +// Reads the byte sitting at bit offset shift of a 64-bit word +template +BOOST_INT128_HOST_DEVICE constexpr ByteType extract_byte(const std::uint64_t word, const unsigned shift) noexcept +{ + return static_cast(static_cast((word >> shift) & UINT64_C(0xFF))); +} + +// Places a byte at bit offset shift of a 64-bit word +template +BOOST_INT128_HOST_DEVICE constexpr std::uint64_t insert_byte(const ByteType value, const unsigned shift) noexcept +{ + return static_cast(static_cast(value)) << shift; +} + +// Rebuilds either library type from the raw two's complement words +template +struct word_builder; + +template <> +struct word_builder +{ + static BOOST_INT128_HOST_DEVICE constexpr uint128 build(const std::uint64_t hi, const std::uint64_t lo) noexcept + { + return uint128{hi, lo}; + } +}; + +template <> +struct word_builder +{ + static BOOST_INT128_HOST_DEVICE constexpr int128 build(const std::uint64_t hi, const std::uint64_t lo) noexcept + { + return from_bits(hi, lo); + } +}; + +// The byte arrays are built from shifts rather than from the object representation, +// so all four functions below are exact on either endianness. Bytes is anything +// that can be indexed with the subscript operator: a std::array or a pointer. + +template +BOOST_INT128_HOST_DEVICE constexpr std::array to_be_bytes_impl(const std::uint64_t hi, const std::uint64_t lo) noexcept +{ + return {{extract_byte(hi, 56U), extract_byte(hi, 48U), + extract_byte(hi, 40U), extract_byte(hi, 32U), + extract_byte(hi, 24U), extract_byte(hi, 16U), + extract_byte(hi, 8U), extract_byte(hi, 0U), + extract_byte(lo, 56U), extract_byte(lo, 48U), + extract_byte(lo, 40U), extract_byte(lo, 32U), + extract_byte(lo, 24U), extract_byte(lo, 16U), + extract_byte(lo, 8U), extract_byte(lo, 0U)}}; +} + +template +BOOST_INT128_HOST_DEVICE constexpr std::array to_le_bytes_impl(const std::uint64_t hi, const std::uint64_t lo) noexcept +{ + return {{extract_byte(lo, 0U), extract_byte(lo, 8U), + extract_byte(lo, 16U), extract_byte(lo, 24U), + extract_byte(lo, 32U), extract_byte(lo, 40U), + extract_byte(lo, 48U), extract_byte(lo, 56U), + extract_byte(hi, 0U), extract_byte(hi, 8U), + extract_byte(hi, 16U), extract_byte(hi, 24U), + extract_byte(hi, 32U), extract_byte(hi, 40U), + extract_byte(hi, 48U), extract_byte(hi, 56U)}}; +} + +template +BOOST_INT128_HOST_DEVICE constexpr T from_be_bytes_impl(const Bytes& bytes) noexcept +{ + return word_builder::build(insert_byte(bytes[0], 56U) | insert_byte(bytes[1], 48U) | + insert_byte(bytes[2], 40U) | insert_byte(bytes[3], 32U) | + insert_byte(bytes[4], 24U) | insert_byte(bytes[5], 16U) | + insert_byte(bytes[6], 8U) | insert_byte(bytes[7], 0U), + insert_byte(bytes[8], 56U) | insert_byte(bytes[9], 48U) | + insert_byte(bytes[10], 40U) | insert_byte(bytes[11], 32U) | + insert_byte(bytes[12], 24U) | insert_byte(bytes[13], 16U) | + insert_byte(bytes[14], 8U) | insert_byte(bytes[15], 0U)); +} + +template +BOOST_INT128_HOST_DEVICE constexpr T from_le_bytes_impl(const Bytes& bytes) noexcept +{ + return word_builder::build(insert_byte(bytes[8], 0U) | insert_byte(bytes[9], 8U) | + insert_byte(bytes[10], 16U) | insert_byte(bytes[11], 24U) | + insert_byte(bytes[12], 32U) | insert_byte(bytes[13], 40U) | + insert_byte(bytes[14], 48U) | insert_byte(bytes[15], 56U), + insert_byte(bytes[0], 0U) | insert_byte(bytes[1], 8U) | + insert_byte(bytes[2], 16U) | insert_byte(bytes[3], 24U) | + insert_byte(bytes[4], 32U) | insert_byte(bytes[5], 40U) | + insert_byte(bytes[6], 48U) | insert_byte(bytes[7], 56U)); +} + +} // namespace detail + +//===================================== +// Byte array conversions +//===================================== + +BOOST_INT128_EXPORT template , bool> = true> +BOOST_INT128_HOST_DEVICE constexpr std::array to_be_bytes(const uint128 value) noexcept +{ + return detail::to_be_bytes_impl(value.high, value.low); +} + +BOOST_INT128_EXPORT template , bool> = true> +BOOST_INT128_HOST_DEVICE constexpr std::array to_be_bytes(const int128 value) noexcept +{ + return detail::to_be_bytes_impl(value.high, value.low); +} + +BOOST_INT128_EXPORT template +BOOST_INT128_HOST_DEVICE constexpr T from_be_bytes(const std::array& bytes) noexcept +{ + static_assert(detail::is_valid_overload_v, + "The target type must be boost::int128::uint128 or boost::int128::int128"); + static_assert(detail::is_byte_like_v, + "The source bytes must be char, signed char, unsigned char, or std::byte"); + static_assert(N == sizeof(T), "The number of bytes provided, and the target type number of bytes do not match"); + + return detail::from_be_bytes_impl(bytes); +} + +// Reads sizeof(T) bytes starting at bytes +BOOST_INT128_EXPORT template +BOOST_INT128_HOST_DEVICE constexpr T from_be_bytes(const ByteType* bytes) noexcept +{ + static_assert(detail::is_valid_overload_v, + "The target type must be boost::int128::uint128 or boost::int128::int128"); + static_assert(detail::is_byte_like_v, + "The source bytes must be char, signed char, unsigned char, or std::byte"); + + return detail::from_be_bytes_impl(bytes); +} + +BOOST_INT128_EXPORT template , bool> = true> +BOOST_INT128_HOST_DEVICE constexpr std::array to_le_bytes(const uint128 value) noexcept +{ + return detail::to_le_bytes_impl(value.high, value.low); +} + +BOOST_INT128_EXPORT template , bool> = true> +BOOST_INT128_HOST_DEVICE constexpr std::array to_le_bytes(const int128 value) noexcept +{ + return detail::to_le_bytes_impl(value.high, value.low); +} + +BOOST_INT128_EXPORT template +BOOST_INT128_HOST_DEVICE constexpr T from_le_bytes(const std::array& bytes) noexcept +{ + static_assert(detail::is_valid_overload_v, + "The target type must be boost::int128::uint128 or boost::int128::int128"); + static_assert(detail::is_byte_like_v, + "The source bytes must be char, signed char, unsigned char, or std::byte"); + static_assert(N == sizeof(T), "The number of bytes provided, and the target type number of bytes do not match"); + + return detail::from_le_bytes_impl(bytes); +} + +// Reads sizeof(T) bytes starting at bytes +BOOST_INT128_EXPORT template +BOOST_INT128_HOST_DEVICE constexpr T from_le_bytes(const ByteType* bytes) noexcept +{ + static_assert(detail::is_valid_overload_v, + "The target type must be boost::int128::uint128 or boost::int128::int128"); + static_assert(detail::is_byte_like_v, + "The source bytes must be char, signed char, unsigned char, or std::byte"); + + return detail::from_le_bytes_impl(bytes); +} + +BOOST_INT128_EXPORT template , bool> = true> +BOOST_INT128_HOST_DEVICE constexpr std::array to_ne_bytes(const uint128 value) noexcept +{ + #if BOOST_INT128_ENDIAN_LITTLE_BYTE + + return detail::to_le_bytes_impl(value.high, value.low); + + #else + + return detail::to_be_bytes_impl(value.high, value.low); + + #endif +} + +BOOST_INT128_EXPORT template , bool> = true> +BOOST_INT128_HOST_DEVICE constexpr std::array to_ne_bytes(const int128 value) noexcept +{ + #if BOOST_INT128_ENDIAN_LITTLE_BYTE + + return detail::to_le_bytes_impl(value.high, value.low); + + #else + + return detail::to_be_bytes_impl(value.high, value.low); + + #endif +} + +BOOST_INT128_EXPORT template +BOOST_INT128_HOST_DEVICE constexpr T from_ne_bytes(const std::array& bytes) noexcept +{ + static_assert(detail::is_valid_overload_v, + "The target type must be boost::int128::uint128 or boost::int128::int128"); + static_assert(detail::is_byte_like_v, + "The source bytes must be char, signed char, unsigned char, or std::byte"); + static_assert(N == sizeof(T), "The number of bytes provided, and the target type number of bytes do not match"); + + #if BOOST_INT128_ENDIAN_LITTLE_BYTE + + return detail::from_le_bytes_impl(bytes); + + #else + + return detail::from_be_bytes_impl(bytes); + + #endif +} + +// Reads sizeof(T) bytes starting at bytes +BOOST_INT128_EXPORT template +BOOST_INT128_HOST_DEVICE constexpr T from_ne_bytes(const ByteType* bytes) noexcept +{ + static_assert(detail::is_valid_overload_v, + "The target type must be boost::int128::uint128 or boost::int128::int128"); + static_assert(detail::is_byte_like_v, + "The source bytes must be char, signed char, unsigned char, or std::byte"); + + #if BOOST_INT128_ENDIAN_LITTLE_BYTE + + return detail::from_le_bytes_impl(bytes); + + #else + + return detail::from_be_bytes_impl(bytes); + + #endif +} + +} // namespace int128 +} // namespace boost + +#endif // BOOST_INT128_BYTE_CONVERSIONS_HPP From 159db0e13da2f2727953b88d2448715508a7a733 Mon Sep 17 00:00:00 2001 From: Matt Borland Date: Mon, 3 Aug 2026 11:35:15 -0400 Subject: [PATCH 2/5] Add testing --- test/Jamfile | 5 + .../byte_conversions_compile.cpp | 10 + .../byte_conversions_extent_fail.cpp | 17 + .../byte_conversions_target_fail.cpp | 17 + test/cuda_jamfile | 5 + test/sycl_jamfile | 5 + test/sycl_test.hpp | 153 +++++++++ test/test_byte_conversions.cpp | 323 ++++++++++++++++++ test/test_signed_byte_arrays.cu | 147 ++++++++ test/test_signed_byte_arrays_sycl.cpp | 12 + test/test_signed_byte_order.cu | 98 ++++++ test/test_signed_byte_order_sycl.cpp | 12 + test/test_unsigned_byte_arrays.cu | 147 ++++++++ test/test_unsigned_byte_arrays_sycl.cpp | 12 + test/test_unsigned_byte_order.cu | 98 ++++++ test/test_unsigned_byte_order_sycl.cpp | 12 + 16 files changed, 1073 insertions(+) create mode 100644 test/compile_tests/byte_conversions_compile.cpp create mode 100644 test/compile_tests/byte_conversions_extent_fail.cpp create mode 100644 test/compile_tests/byte_conversions_target_fail.cpp create mode 100644 test/test_byte_conversions.cpp create mode 100644 test/test_signed_byte_arrays.cu create mode 100644 test/test_signed_byte_arrays_sycl.cpp create mode 100644 test/test_signed_byte_order.cu create mode 100644 test/test_signed_byte_order_sycl.cpp create mode 100644 test/test_unsigned_byte_arrays.cu create mode 100644 test/test_unsigned_byte_arrays_sycl.cpp create mode 100644 test/test_unsigned_byte_order.cu create mode 100644 test/test_unsigned_byte_order_sycl.cpp diff --git a/test/Jamfile b/test/Jamfile index 93192cc4..b8c226b0 100644 --- a/test/Jamfile +++ b/test/Jamfile @@ -71,6 +71,7 @@ run test_limits_i128.cpp ; run test_climits.cpp ; run test_bit.cpp ; +run test_byte_conversions.cpp ; run test_literals.cpp ; run test_from_chars_bases.cpp ; run test_stream.cpp ; @@ -118,6 +119,7 @@ run test_layout.cpp ; run ../examples/hello_world.cpp ; run ../examples/construction.cpp ; run ../examples/bit.cpp ; +run ../examples/byte_conversions.cpp ; run ../examples/saturating_arithmetic.cpp ; run ../examples/checked_arithmetic.cpp ; run ../examples/integer_comparison.cpp ; @@ -153,6 +155,7 @@ compile github_issue_480.cpp ; # Compilation of individual headers compile compile_tests/int128_master_header_compile.cpp ; compile compile_tests/bit_compile.cpp ; +compile compile_tests/byte_conversions_compile.cpp ; compile compile_tests/fmt_format_compile.cpp ; compile compile_tests/charconv_compile.cpp ; compile compile_tests/climits_compile.cpp ; @@ -172,3 +175,5 @@ compile-fail compile_tests/literals_overflow_fail.cpp ; compile-fail compile_tests/literals_base_prefix_fail.cpp ; compile-fail compile_tests/literals_invalid_fail.cpp ; compile-fail compile_tests/utilities_comparison_fail.cpp ; +compile-fail compile_tests/byte_conversions_extent_fail.cpp ; +compile-fail compile_tests/byte_conversions_target_fail.cpp ; diff --git a/test/compile_tests/byte_conversions_compile.cpp b/test/compile_tests/byte_conversions_compile.cpp new file mode 100644 index 00000000..ae20a4ed --- /dev/null +++ b/test/compile_tests/byte_conversions_compile.cpp @@ -0,0 +1,10 @@ +// Copyright 2026 Matt Borland +// Distributed under the Boost Software License, Version 1.0. +// https://www.boost.org/LICENSE_1_0.txt + +#include + +int main() +{ + return 0; +} diff --git a/test/compile_tests/byte_conversions_extent_fail.cpp b/test/compile_tests/byte_conversions_extent_fail.cpp new file mode 100644 index 00000000..2b8d4560 --- /dev/null +++ b/test/compile_tests/byte_conversions_extent_fail.cpp @@ -0,0 +1,17 @@ +// Copyright 2026 Matt Borland +// Distributed under the Boost Software License, Version 1.0. +// https://www.boost.org/LICENSE_1_0.txt + +// A byte array whose size does not match the width of the target type is +// rejected at compile time. This must fail to compile. + +#include +#include +#include + +int main() +{ + const std::array bytes {}; + + return static_cast(boost::int128::from_be_bytes(bytes)); +} diff --git a/test/compile_tests/byte_conversions_target_fail.cpp b/test/compile_tests/byte_conversions_target_fail.cpp new file mode 100644 index 00000000..52289472 --- /dev/null +++ b/test/compile_tests/byte_conversions_target_fail.cpp @@ -0,0 +1,17 @@ +// Copyright 2026 Matt Borland +// Distributed under the Boost Software License, Version 1.0. +// https://www.boost.org/LICENSE_1_0.txt + +// The target of from_be_bytes must be one of the library types, so a built-in +// integer is rejected. This must fail to compile. + +#include +#include +#include + +int main() +{ + const std::array bytes {}; + + return static_cast(boost::int128::from_be_bytes(bytes)); +} diff --git a/test/cuda_jamfile b/test/cuda_jamfile index fbd26a4f..abf9250a 100644 --- a/test/cuda_jamfile +++ b/test/cuda_jamfile @@ -46,6 +46,11 @@ run test_rotr.cu ; run test_popcount.cu ; run test_byteswap.cu ; +run test_unsigned_byte_order.cu ; +run test_signed_byte_order.cu ; +run test_unsigned_byte_arrays.cu ; +run test_signed_byte_arrays.cu ; + run test_unsigned_eq.cu ; run test_signed_eq.cu ; run test_unsigned_ne.cu ; diff --git a/test/sycl_jamfile b/test/sycl_jamfile index 046fb068..2aeb4fca 100644 --- a/test/sycl_jamfile +++ b/test/sycl_jamfile @@ -46,6 +46,11 @@ run test_rotr_sycl.cpp ; run test_popcount_sycl.cpp ; run test_byteswap_sycl.cpp ; +run test_unsigned_byte_order_sycl.cpp ; +run test_signed_byte_order_sycl.cpp ; +run test_unsigned_byte_arrays_sycl.cpp ; +run test_signed_byte_arrays_sycl.cpp ; + run test_unsigned_eq_sycl.cpp ; run test_signed_eq_sycl.cpp ; run test_unsigned_ne_sycl.cpp ; diff --git a/test/sycl_test.hpp b/test/sycl_test.hpp index b2157108..50f3edd8 100644 --- a/test/sycl_test.hpp +++ b/test/sycl_test.hpp @@ -12,6 +12,7 @@ #include #include #include +#include #include #include #include @@ -181,6 +182,158 @@ int run_compare(Pred pred) return EXIT_FAILURE; } +// Byte order runner. Converts each value to big-endian and to little-endian order on the +// device, and also runs a full round trip, which has to recover the input. +template +int run_byte_order() +{ + sycl::queue q; + std::cout << "SYCL device: " + << q.get_device().get_info() << "\n"; + + T* in {sycl::malloc_shared(num_elements, q)}; + T* out_be {sycl::malloc_shared(num_elements, q)}; + T* out_le {sycl::malloc_shared(num_elements, q)}; + T* out_round {sycl::malloc_shared(num_elements, q)}; + + std::mt19937_64 rng {42}; + for (int i {0}; i < num_elements; ++i) + { + in[i] = random_value(rng); + } + + q.submit([&](sycl::handler& h) + { + h.parallel_for(sycl::range<1>(num_elements), [=](sycl::id<1> idx) + { + const int i {static_cast(idx[0])}; + out_be[i] = boost::int128::to_be(in[i]); + out_le[i] = boost::int128::to_le(in[i]); + out_round[i] = boost::int128::from_le(boost::int128::to_le(in[i])); + }); + }).wait(); + + int failures {0}; + for (int i {0}; i < num_elements; ++i) + { + if (out_be[i] != boost::int128::to_be(in[i]) || + out_le[i] != boost::int128::to_le(in[i]) || + out_round[i] != in[i]) + { + if (failures < 5) + { + std::cerr << "Mismatch at element " << i << "\n"; + } + ++failures; + } + } + + sycl::free(in, q); + sycl::free(out_be, q); + sycl::free(out_le, q); + sycl::free(out_round, q); + + if (failures == 0) + { + std::cout << "Test PASSED\n"; + return EXIT_SUCCESS; + } + + std::cerr << "Test FAILED with " << failures << " mismatches\n"; + return EXIT_FAILURE; +} + +// Byte array runner. Writes each value out as bytes in all three orders on the device, and +// reads a separate random byte buffer back as values, comparing everything with the host. +// The arrays are written by assignment and read through a raw pointer so that no +// std::array member function is called from device code. +template +int run_byte_arrays() +{ + using bytes_type = std::array; + + sycl::queue q; + std::cout << "SYCL device: " + << q.get_device().get_info() << "\n"; + + T* in {sycl::malloc_shared(num_elements, q)}; + std::uint8_t* in_bytes {sycl::malloc_shared(num_elements * sizeof(T), q)}; + bytes_type* out_be {sycl::malloc_shared(num_elements, q)}; + bytes_type* out_le {sycl::malloc_shared(num_elements, q)}; + bytes_type* out_ne {sycl::malloc_shared(num_elements, q)}; + T* from_be {sycl::malloc_shared(num_elements, q)}; + T* from_le {sycl::malloc_shared(num_elements, q)}; + T* from_ne {sycl::malloc_shared(num_elements, q)}; + + std::mt19937_64 rng {42}; + for (int i {0}; i < num_elements; ++i) + { + in[i] = random_value(rng); + } + for (std::size_t i {0}; i < num_elements * sizeof(T); ++i) + { + in_bytes[i] = static_cast(rng() & 0xFFU); + } + + q.submit([&](sycl::handler& h) + { + h.parallel_for(sycl::range<1>(num_elements), [=](sycl::id<1> idx) + { + const int i {static_cast(idx[0])}; + + out_be[i] = boost::int128::to_be_bytes(in[i]); + out_le[i] = boost::int128::to_le_bytes(in[i]); + out_ne[i] = boost::int128::to_ne_bytes(in[i]); + + const std::uint8_t* bytes {in_bytes + static_cast(i) * sizeof(T)}; + from_be[i] = boost::int128::from_be_bytes(bytes); + from_le[i] = boost::int128::from_le_bytes(bytes); + from_ne[i] = boost::int128::from_ne_bytes(bytes); + }); + }).wait(); + + int failures {0}; + for (int i {0}; i < num_elements; ++i) + { + const std::uint8_t* bytes {in_bytes + static_cast(i) * sizeof(T)}; + + if (out_be[i] != boost::int128::to_be_bytes(in[i]) || + out_le[i] != boost::int128::to_le_bytes(in[i]) || + out_ne[i] != boost::int128::to_ne_bytes(in[i]) || + from_be[i] != boost::int128::from_be_bytes(bytes) || + from_le[i] != boost::int128::from_le_bytes(bytes) || + from_ne[i] != boost::int128::from_ne_bytes(bytes) || + boost::int128::from_be_bytes(out_be[i]) != in[i] || + boost::int128::from_le_bytes(out_le[i]) != in[i] || + boost::int128::from_ne_bytes(out_ne[i]) != in[i]) + { + if (failures < 5) + { + std::cerr << "Mismatch at element " << i << "\n"; + } + ++failures; + } + } + + sycl::free(in, q); + sycl::free(in_bytes, q); + sycl::free(out_be, q); + sycl::free(out_le, q); + sycl::free(out_ne, q); + sycl::free(from_be, q); + sycl::free(from_le, q); + sycl::free(from_ne, q); + + if (failures == 0) + { + std::cout << "Test PASSED\n"; + return EXIT_SUCCESS; + } + + std::cerr << "Test FAILED with " << failures << " mismatches\n"; + return EXIT_FAILURE; +} + // to_chars runner. Formats each value on the device via boost::charconv::to_chars (now // SYCL device-capable) and compares the produced text against the host. template diff --git a/test/test_byte_conversions.cpp b/test/test_byte_conversions.cpp new file mode 100644 index 00000000..ce35bc8f --- /dev/null +++ b/test/test_byte_conversions.cpp @@ -0,0 +1,323 @@ +// Copyright 2026 Matt Borland +// Distributed under the Boost Software License, Version 1.0. +// https://www.boost.org/LICENSE_1_0.txt + +#ifndef BOOST_INT128_BUILD_MODULE + +#include +#include +#include +#include +#include + +#else + +import boost.int128; + +#endif + +#include +#include +#include +#include +#include +#include + +using boost::int128::uint128; +using boost::int128::int128; + +static const uint128 pattern_u {UINT64_C(0x0102030405060708), UINT64_C(0x090A0B0C0D0E0F10)}; + +// The endian macros are not part of the module interface, so the native order is +// taken from the object representation of a known value instead +static bool native_is_little() +{ + const uint128 probe {UINT64_C(0), UINT64_C(1)}; + std::uint8_t raw[sizeof(uint128)] {}; + std::memcpy(raw, &probe, sizeof(uint128)); + + return raw[0] == 1U; +} + +template +void check_bytes(const std::array& actual, const std::array& expected) +{ + for (std::size_t i {}; i < actual.size(); ++i) + { + BOOST_TEST_EQ(static_cast(actual[i]), expected[i]); + } +} + +// SplitMix64 keeps the test free of any dependency +static std::uint64_t next_word(std::uint64_t& state) +{ + state += UINT64_C(0x9E3779B97F4A7C15); + auto z {state}; + z = (z ^ (z >> 30U)) * UINT64_C(0xBF58476D1CE4E5B9); + z = (z ^ (z >> 27U)) * UINT64_C(0x94D049BB133111EB); + + return z ^ (z >> 31U); +} + +void test_to_be_bytes_u128() +{ + check_bytes(boost::int128::to_be_bytes(pattern_u), + {{0x01, 0x02, 0x03, 0x04, 0x05, 0x06, 0x07, 0x08, 0x09, 0x0A, 0x0B, 0x0C, 0x0D, 0x0E, 0x0F, 0x10}}); + + check_bytes(boost::int128::to_be_bytes(uint128{0}), + {{0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0}}); + + check_bytes(boost::int128::to_be_bytes(uint128{42}), + {{0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0x2A}}); + + check_bytes(boost::int128::to_be_bytes((std::numeric_limits::max)()), + {{0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF}}); +} + +void test_to_le_bytes_u128() +{ + check_bytes(boost::int128::to_le_bytes(pattern_u), + {{0x10, 0x0F, 0x0E, 0x0D, 0x0C, 0x0B, 0x0A, 0x09, 0x08, 0x07, 0x06, 0x05, 0x04, 0x03, 0x02, 0x01}}); + + check_bytes(boost::int128::to_le_bytes(uint128{42}), + {{0x2A, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0}}); + + check_bytes(boost::int128::to_le_bytes(uint128{UINT64_C(1), UINT64_C(0)}), + {{0, 0, 0, 0, 0, 0, 0, 0, 0x01, 0, 0, 0, 0, 0, 0, 0}}); +} + +void test_to_be_bytes_i128() +{ + check_bytes(boost::int128::to_be_bytes(int128{-1}), + {{0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF}}); + + check_bytes(boost::int128::to_be_bytes(int128{-42}), + {{0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xD6}}); + + // The minimum is the sign bit alone, and the maximum is everything but the sign bit + check_bytes(boost::int128::to_be_bytes((std::numeric_limits::min)()), + {{0x80, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0}}); + + check_bytes(boost::int128::to_be_bytes((std::numeric_limits::max)()), + {{0x7F, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF}}); +} + +void test_to_le_bytes_i128() +{ + check_bytes(boost::int128::to_le_bytes(int128{-42}), + {{0xD6, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF}}); + + check_bytes(boost::int128::to_le_bytes((std::numeric_limits::min)()), + {{0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0x80}}); +} + +void test_from_be_bytes() +{ + const std::array bytes + {{0x01, 0x02, 0x03, 0x04, 0x05, 0x06, 0x07, 0x08, 0x09, 0x0A, 0x0B, 0x0C, 0x0D, 0x0E, 0x0F, 0x10}}; + + BOOST_TEST_EQ(boost::int128::from_be_bytes(bytes), pattern_u); + BOOST_TEST_EQ(boost::int128::from_be_bytes(bytes), static_cast(pattern_u)); + + // The same bytes read through the pointer overload + BOOST_TEST_EQ(boost::int128::from_be_bytes(bytes.data()), pattern_u); + + const std::array negative + {{0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xD6}}; + + BOOST_TEST_EQ(boost::int128::from_be_bytes(negative), int128{-42}); + BOOST_TEST_EQ(boost::int128::from_be_bytes(negative), (std::numeric_limits::max)() - 41U); + + const std::array zero {}; + BOOST_TEST_EQ(boost::int128::from_be_bytes(zero), uint128{0}); + BOOST_TEST_EQ(boost::int128::from_be_bytes(zero), int128{0}); +} + +void test_from_le_bytes() +{ + const std::array bytes + {{0x10, 0x0F, 0x0E, 0x0D, 0x0C, 0x0B, 0x0A, 0x09, 0x08, 0x07, 0x06, 0x05, 0x04, 0x03, 0x02, 0x01}}; + + BOOST_TEST_EQ(boost::int128::from_le_bytes(bytes), pattern_u); + BOOST_TEST_EQ(boost::int128::from_le_bytes(bytes), static_cast(pattern_u)); + BOOST_TEST_EQ(boost::int128::from_le_bytes(bytes.data()), pattern_u); + + const std::array negative + {{0xD6, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF, 0xFF}}; + + BOOST_TEST_EQ(boost::int128::from_le_bytes(negative), int128{-42}); +} + +void test_ne_bytes() +{ + const auto ne {boost::int128::to_ne_bytes(pattern_u)}; + + // Native order is by definition the object representation + std::uint8_t raw[sizeof(uint128)] {}; + std::memcpy(raw, &pattern_u, sizeof(uint128)); + BOOST_TEST_EQ(std::memcmp(raw, ne.data(), sizeof(uint128)), 0); + + if (native_is_little()) + { + BOOST_TEST(ne == boost::int128::to_le_bytes(pattern_u)); + } + else + { + BOOST_TEST(ne == boost::int128::to_be_bytes(pattern_u)); + } + + BOOST_TEST_EQ(boost::int128::from_ne_bytes(ne), pattern_u); + BOOST_TEST_EQ(boost::int128::from_ne_bytes(ne.data()), pattern_u); + + const auto signed_ne {boost::int128::to_ne_bytes(int128{-42})}; + BOOST_TEST_EQ(boost::int128::from_ne_bytes(signed_ne), int128{-42}); + BOOST_TEST_EQ(boost::int128::from_ne_bytes(signed_ne.data()), int128{-42}); +} + +void test_scalar_swaps() +{ + // On the matching platform the conversion is the identity, and a byteswap otherwise + if (native_is_little()) + { + BOOST_TEST_EQ(boost::int128::to_le(pattern_u), pattern_u); + BOOST_TEST_EQ(boost::int128::to_be(pattern_u), boost::int128::byteswap(pattern_u)); + } + else + { + BOOST_TEST_EQ(boost::int128::to_be(pattern_u), pattern_u); + BOOST_TEST_EQ(boost::int128::to_le(pattern_u), boost::int128::byteswap(pattern_u)); + } + + // The big-endian image of a value always reads back as the big-endian byte array + check_bytes(boost::int128::to_ne_bytes(boost::int128::to_be(pattern_u)), + {{0x01, 0x02, 0x03, 0x04, 0x05, 0x06, 0x07, 0x08, 0x09, 0x0A, 0x0B, 0x0C, 0x0D, 0x0E, 0x0F, 0x10}}); + + check_bytes(boost::int128::to_ne_bytes(boost::int128::to_le(pattern_u)), + {{0x10, 0x0F, 0x0E, 0x0D, 0x0C, 0x0B, 0x0A, 0x09, 0x08, 0x07, 0x06, 0x05, 0x04, 0x03, 0x02, 0x01}}); + + std::uint64_t state {UINT64_C(0xDEADBEEFCAFEF00D)}; + + for (int i {}; i < 1024; ++i) + { + const uint128 unsigned_value {next_word(state), next_word(state)}; + const auto signed_value {static_cast(unsigned_value)}; + + BOOST_TEST_EQ(boost::int128::from_be(boost::int128::to_be(unsigned_value)), unsigned_value); + BOOST_TEST_EQ(boost::int128::from_le(boost::int128::to_le(unsigned_value)), unsigned_value); + BOOST_TEST_EQ(boost::int128::from_be(boost::int128::to_be(signed_value)), signed_value); + BOOST_TEST_EQ(boost::int128::from_le(boost::int128::to_le(signed_value)), signed_value); + + // to_be and to_le differ by exactly a byteswap + BOOST_TEST_EQ(boost::int128::to_be(unsigned_value), boost::int128::byteswap(boost::int128::to_le(unsigned_value))); + } +} + +template +void test_array_round_trip(const T value) +{ + const auto be {boost::int128::to_be_bytes(value)}; + const auto le {boost::int128::to_le_bytes(value)}; + const auto ne {boost::int128::to_ne_bytes(value)}; + + BOOST_TEST_EQ(boost::int128::from_be_bytes(be), value); + BOOST_TEST_EQ(boost::int128::from_le_bytes(le), value); + BOOST_TEST_EQ(boost::int128::from_ne_bytes(ne), value); + + BOOST_TEST_EQ(boost::int128::from_be_bytes(be.data()), value); + BOOST_TEST_EQ(boost::int128::from_le_bytes(le.data()), value); + BOOST_TEST_EQ(boost::int128::from_ne_bytes(ne.data()), value); + + // One order is the reverse of the other + for (std::size_t i {}; i < be.size(); ++i) + { + BOOST_TEST_EQ(be[i], le[be.size() - 1U - i]); + } +} + +void test_array_round_trips() +{ + test_array_round_trip(uint128{0}); + test_array_round_trip(uint128{1}); + test_array_round_trip(pattern_u); + test_array_round_trip((std::numeric_limits::max)()); + + test_array_round_trip(int128{0}); + test_array_round_trip(int128{-1}); + test_array_round_trip(int128{-42}); + test_array_round_trip((std::numeric_limits::min)()); + test_array_round_trip((std::numeric_limits::max)()); + + std::uint64_t state {UINT64_C(0x0123456789ABCDEF)}; + + for (int i {}; i < 1024; ++i) + { + const uint128 unsigned_value {next_word(state), next_word(state)}; + test_array_round_trip(unsigned_value); + test_array_round_trip(static_cast(unsigned_value)); + } +} + +void test_element_types() +{ + const auto as_char {boost::int128::to_be_bytes(pattern_u)}; + BOOST_TEST_EQ(boost::int128::from_be_bytes(as_char), pattern_u); + BOOST_TEST_EQ(boost::int128::from_be_bytes(as_char.data()), pattern_u); + + const auto as_signed_char {boost::int128::to_le_bytes(int128{-42})}; + BOOST_TEST_EQ(boost::int128::from_le_bytes(as_signed_char), int128{-42}); + + const auto as_unsigned_char {boost::int128::to_ne_bytes(pattern_u)}; + BOOST_TEST_EQ(boost::int128::from_ne_bytes(as_unsigned_char), pattern_u); + + #if defined(__cpp_lib_byte) && __cpp_lib_byte >= 201603L + + const auto as_byte {boost::int128::to_be_bytes(pattern_u)}; + BOOST_TEST(as_byte[0] == std::byte{0x01}); + BOOST_TEST(as_byte[15] == std::byte{0x10}); + BOOST_TEST_EQ(boost::int128::from_be_bytes(as_byte), pattern_u); + BOOST_TEST_EQ(boost::int128::from_be_bytes(as_byte.data()), pattern_u); + + #endif +} + +void test_constexpr() +{ + constexpr uint128 value {UINT64_C(0x0102030405060708), UINT64_C(0x090A0B0C0D0E0F10)}; + + static_assert(boost::int128::from_be(boost::int128::to_be(value)) == value, "Scalar round trip"); + static_assert(boost::int128::from_le(boost::int128::to_le(value)) == value, "Scalar round trip"); + + constexpr auto be {boost::int128::to_be_bytes(value)}; + static_assert(be[0] == 0x01 && be[15] == 0x10, "Big-endian order"); + static_assert(boost::int128::from_be_bytes(be) == value, "Array round trip"); + + constexpr auto le {boost::int128::to_le_bytes(value)}; + static_assert(le[0] == 0x10 && le[15] == 0x01, "Little-endian order"); + static_assert(boost::int128::from_le_bytes(le) == value, "Array round trip"); + + constexpr auto ne {boost::int128::to_ne_bytes(value)}; + static_assert(boost::int128::from_ne_bytes(ne) == value, "Array round trip"); + + constexpr int128 negative {-42}; + static_assert(boost::int128::from_be_bytes(boost::int128::to_be_bytes(negative)) == negative, "Signed array round trip"); + static_assert(boost::int128::from_le_bytes(boost::int128::to_le_bytes(negative)) == negative, "Signed array round trip"); +} + +int main() +{ + test_to_be_bytes_u128(); + test_to_le_bytes_u128(); + test_to_be_bytes_i128(); + test_to_le_bytes_i128(); + + test_from_be_bytes(); + test_from_le_bytes(); + + test_ne_bytes(); + test_scalar_swaps(); + test_array_round_trips(); + test_element_types(); + test_constexpr(); + + return boost::report_errors(); +} diff --git a/test/test_signed_byte_arrays.cu b/test/test_signed_byte_arrays.cu new file mode 100644 index 00000000..72f6a5f0 --- /dev/null +++ b/test/test_signed_byte_arrays.cu @@ -0,0 +1,147 @@ +// Copyright Matt Borland 2026. +// Use, modification and distribution are subject to the +// Boost Software License, Version 1.0. (See accompanying file +// LICENSE_1_0.txt or copy at http://www.boost.org/LICENSE_1_0.txt) + +#include +#include +#include +#include +#include +#include +#include +#include +#include +#include +#include "cuda_managed_ptr.hpp" +#include "stopwatch.hpp" + +#include + +using test_type = boost::int128::int128; +using bytes_type = std::array; + +// The byte arrays are written by assignment and read through a raw pointer, so no +// std::array member function is called from device code +__global__ void cuda_test(const test_type *in, const std::uint8_t *in_bytes, + bytes_type *out_be, bytes_type *out_le, bytes_type *out_ne, + test_type *from_be, test_type *from_le, test_type *from_ne, int numElements) +{ + int i = blockDim.x * blockIdx.x + threadIdx.x; + + if (i < numElements) + { + out_be[i] = boost::int128::to_be_bytes(in[i]); + out_le[i] = boost::int128::to_le_bytes(in[i]); + out_ne[i] = boost::int128::to_ne_bytes(in[i]); + + const std::uint8_t *bytes = in_bytes + i * static_cast(sizeof(test_type)); + from_be[i] = boost::int128::from_be_bytes(bytes); + from_le[i] = boost::int128::from_le_bytes(bytes); + from_ne[i] = boost::int128::from_ne_bytes(bytes); + } +} + +int main(void) +{ + std::mt19937_64 rng {42}; + + cudaError_t err = cudaSuccess; + + int numElements = 50000; + std::cout << "[Vector operation on " << numElements << " elements]" << std::endl; + + cuda_managed_ptr input_vector(numElements); + cuda_managed_ptr input_bytes(numElements * sizeof(test_type)); + cuda_managed_ptr output_be(numElements); + cuda_managed_ptr output_le(numElements); + cuda_managed_ptr output_ne(numElements); + cuda_managed_ptr output_from_be(numElements); + cuda_managed_ptr output_from_le(numElements); + cuda_managed_ptr output_from_ne(numElements); + + boost::random::uniform_int_distribution dist {(std::numeric_limits::min)(), (std::numeric_limits::max)()}; + boost::random::uniform_int_distribution byte_dist {0U, 255U}; + for (std::size_t i = 0; i < numElements; ++i) + { + input_vector[i] = dist(rng); + } + for (std::size_t i = 0; i < numElements * sizeof(test_type); ++i) + { + input_bytes[i] = static_cast(byte_dist(rng)); + } + + int threadsPerBlock = 256; + int blocksPerGrid = (numElements + threadsPerBlock - 1) / threadsPerBlock; + std::cout << "CUDA kernel launch with " << blocksPerGrid << " blocks of " << threadsPerBlock << " threads" << std::endl; + + watch w; + + cuda_test<<>>(input_vector.get(), input_bytes.get(), + output_be.get(), output_le.get(), output_ne.get(), + output_from_be.get(), output_from_le.get(), output_from_ne.get(), + numElements); + cudaDeviceSynchronize(); + + std::cout << "CUDA kernal done in: " << w.elapsed() << "s" << std::endl; + + err = cudaGetLastError(); + if (err != cudaSuccess) + { + std::cerr << "Failed to launch kernel (error code " << cudaGetErrorString(err) << ")!" << std::endl; + return EXIT_FAILURE; + } + + const std::uint8_t *host_bytes = input_bytes.get(); + + w.reset(); + std::vector be_results; + std::vector le_results; + std::vector ne_results; + be_results.reserve(numElements); + le_results.reserve(numElements); + ne_results.reserve(numElements); + for (int i = 0; i < numElements; ++i) + { + be_results.push_back(boost::int128::to_be_bytes(input_vector[i])); + le_results.push_back(boost::int128::to_le_bytes(input_vector[i])); + ne_results.push_back(boost::int128::to_ne_bytes(input_vector[i])); + } + double t = w.elapsed(); + + for (int i = 0; i < numElements; ++i) + { + if (output_be[i] != be_results[i] || output_le[i] != le_results[i] || output_ne[i] != ne_results[i]) + { + std::cerr << "Result verification failed at element " << i << "!" << std::endl; + return EXIT_FAILURE; + } + + const std::uint8_t *bytes = host_bytes + i * static_cast(sizeof(test_type)); + + if (output_from_be[i] != boost::int128::from_be_bytes(bytes) || + output_from_le[i] != boost::int128::from_le_bytes(bytes) || + output_from_ne[i] != boost::int128::from_ne_bytes(bytes)) + { + std::cerr << "Result verification failed at element " << i << "!" << std::endl; + return EXIT_FAILURE; + } + + // Reading back what was written recovers the value. The pointer overload is used + // rather than the std::array one because nvcc instantiates the device side of a + // __host__ __device__ template even for a host call, and the std::array overload + // reads its argument through std::array::operator[], a constexpr host function. + if (boost::int128::from_be_bytes(output_be[i].data()) != input_vector[i] || + boost::int128::from_le_bytes(output_le[i].data()) != input_vector[i] || + boost::int128::from_ne_bytes(output_ne[i].data()) != input_vector[i]) + { + std::cerr << "Round trip verification failed at element " << i << "!" << std::endl; + return EXIT_FAILURE; + } + } + + std::cout << "Test PASSED, normal calculation time: " << t << "s" << std::endl; + std::cout << "Done\n"; + + return 0; +} diff --git a/test/test_signed_byte_arrays_sycl.cpp b/test/test_signed_byte_arrays_sycl.cpp new file mode 100644 index 00000000..90753c4b --- /dev/null +++ b/test/test_signed_byte_arrays_sycl.cpp @@ -0,0 +1,12 @@ +// Copyright 2026 Matt Borland +// Distributed under the Boost Software License, Version 1.0. +// https://www.boost.org/LICENSE_1_0.txt + +#include "sycl_test.hpp" + +using boost::int128::int128; + +int main() +{ + return int128_sycl_test::run_byte_arrays(); +} diff --git a/test/test_signed_byte_order.cu b/test/test_signed_byte_order.cu new file mode 100644 index 00000000..f11d79f6 --- /dev/null +++ b/test/test_signed_byte_order.cu @@ -0,0 +1,98 @@ +// Copyright Matt Borland 2026. +// Use, modification and distribution are subject to the +// Boost Software License, Version 1.0. (See accompanying file +// LICENSE_1_0.txt or copy at http://www.boost.org/LICENSE_1_0.txt) + +#include +#include +#include +#include +#include +#include +#include +#include +#include "cuda_managed_ptr.hpp" +#include "stopwatch.hpp" + +#include + +using test_type = boost::int128::int128; + +__global__ void cuda_test(const test_type *in, test_type *out_be, test_type *out_le, test_type *out_round, int numElements) +{ + int i = blockDim.x * blockIdx.x + threadIdx.x; + + if (i < numElements) + { + out_be[i] = boost::int128::to_be(in[i]); + out_le[i] = boost::int128::to_le(in[i]); + + // Both directions on the device, which must recover the input + out_round[i] = boost::int128::from_be(boost::int128::to_be(in[i])); + } +} + +int main(void) +{ + std::mt19937_64 rng {42}; + + cudaError_t err = cudaSuccess; + + int numElements = 50000; + std::cout << "[Vector operation on " << numElements << " elements]" << std::endl; + + cuda_managed_ptr input_vector(numElements); + cuda_managed_ptr output_be(numElements); + cuda_managed_ptr output_le(numElements); + cuda_managed_ptr output_round(numElements); + + boost::random::uniform_int_distribution dist {(std::numeric_limits::min)(), (std::numeric_limits::max)()}; + for (std::size_t i = 0; i < numElements; ++i) + { + input_vector[i] = dist(rng); + } + + int threadsPerBlock = 256; + int blocksPerGrid = (numElements + threadsPerBlock - 1) / threadsPerBlock; + std::cout << "CUDA kernel launch with " << blocksPerGrid << " blocks of " << threadsPerBlock << " threads" << std::endl; + + watch w; + + cuda_test<<>>(input_vector.get(), output_be.get(), output_le.get(), output_round.get(), numElements); + cudaDeviceSynchronize(); + + std::cout << "CUDA kernal done in: " << w.elapsed() << "s" << std::endl; + + err = cudaGetLastError(); + if (err != cudaSuccess) + { + std::cerr << "Failed to launch kernel (error code " << cudaGetErrorString(err) << ")!" << std::endl; + return EXIT_FAILURE; + } + + std::vector be_results; + std::vector le_results; + be_results.reserve(numElements); + le_results.reserve(numElements); + w.reset(); + for (int i = 0; i < numElements; ++i) + { + be_results.push_back(boost::int128::to_be(input_vector[i])); + le_results.push_back(boost::int128::to_le(input_vector[i])); + } + double t = w.elapsed(); + + for (int i = 0; i < numElements; ++i) + { + if (output_be[i] != be_results[i] || output_le[i] != le_results[i] || output_round[i] != input_vector[i]) + { + std::cerr << "Result verification failed at element " << i << "!" << std::endl; + return EXIT_FAILURE; + } + } + + std::cout << "Test PASSED, normal calculation time: " << t << "s" << std::endl; + std::cout << "Done\n"; + + return 0; +} diff --git a/test/test_signed_byte_order_sycl.cpp b/test/test_signed_byte_order_sycl.cpp new file mode 100644 index 00000000..2e050445 --- /dev/null +++ b/test/test_signed_byte_order_sycl.cpp @@ -0,0 +1,12 @@ +// Copyright 2026 Matt Borland +// Distributed under the Boost Software License, Version 1.0. +// https://www.boost.org/LICENSE_1_0.txt + +#include "sycl_test.hpp" + +using boost::int128::int128; + +int main() +{ + return int128_sycl_test::run_byte_order(); +} diff --git a/test/test_unsigned_byte_arrays.cu b/test/test_unsigned_byte_arrays.cu new file mode 100644 index 00000000..69976c1e --- /dev/null +++ b/test/test_unsigned_byte_arrays.cu @@ -0,0 +1,147 @@ +// Copyright Matt Borland 2026. +// Use, modification and distribution are subject to the +// Boost Software License, Version 1.0. (See accompanying file +// LICENSE_1_0.txt or copy at http://www.boost.org/LICENSE_1_0.txt) + +#include +#include +#include +#include +#include +#include +#include +#include +#include +#include +#include "cuda_managed_ptr.hpp" +#include "stopwatch.hpp" + +#include + +using test_type = boost::int128::uint128; +using bytes_type = std::array; + +// The byte arrays are written by assignment and read through a raw pointer, so no +// std::array member function is called from device code +__global__ void cuda_test(const test_type *in, const std::uint8_t *in_bytes, + bytes_type *out_be, bytes_type *out_le, bytes_type *out_ne, + test_type *from_be, test_type *from_le, test_type *from_ne, int numElements) +{ + int i = blockDim.x * blockIdx.x + threadIdx.x; + + if (i < numElements) + { + out_be[i] = boost::int128::to_be_bytes(in[i]); + out_le[i] = boost::int128::to_le_bytes(in[i]); + out_ne[i] = boost::int128::to_ne_bytes(in[i]); + + const std::uint8_t *bytes = in_bytes + i * static_cast(sizeof(test_type)); + from_be[i] = boost::int128::from_be_bytes(bytes); + from_le[i] = boost::int128::from_le_bytes(bytes); + from_ne[i] = boost::int128::from_ne_bytes(bytes); + } +} + +int main(void) +{ + std::mt19937_64 rng {42}; + + cudaError_t err = cudaSuccess; + + int numElements = 50000; + std::cout << "[Vector operation on " << numElements << " elements]" << std::endl; + + cuda_managed_ptr input_vector(numElements); + cuda_managed_ptr input_bytes(numElements * sizeof(test_type)); + cuda_managed_ptr output_be(numElements); + cuda_managed_ptr output_le(numElements); + cuda_managed_ptr output_ne(numElements); + cuda_managed_ptr output_from_be(numElements); + cuda_managed_ptr output_from_le(numElements); + cuda_managed_ptr output_from_ne(numElements); + + boost::random::uniform_int_distribution dist {test_type{0U}, (std::numeric_limits::max)()}; + boost::random::uniform_int_distribution byte_dist {0U, 255U}; + for (std::size_t i = 0; i < numElements; ++i) + { + input_vector[i] = dist(rng); + } + for (std::size_t i = 0; i < numElements * sizeof(test_type); ++i) + { + input_bytes[i] = static_cast(byte_dist(rng)); + } + + int threadsPerBlock = 256; + int blocksPerGrid = (numElements + threadsPerBlock - 1) / threadsPerBlock; + std::cout << "CUDA kernel launch with " << blocksPerGrid << " blocks of " << threadsPerBlock << " threads" << std::endl; + + watch w; + + cuda_test<<>>(input_vector.get(), input_bytes.get(), + output_be.get(), output_le.get(), output_ne.get(), + output_from_be.get(), output_from_le.get(), output_from_ne.get(), + numElements); + cudaDeviceSynchronize(); + + std::cout << "CUDA kernal done in: " << w.elapsed() << "s" << std::endl; + + err = cudaGetLastError(); + if (err != cudaSuccess) + { + std::cerr << "Failed to launch kernel (error code " << cudaGetErrorString(err) << ")!" << std::endl; + return EXIT_FAILURE; + } + + const std::uint8_t *host_bytes = input_bytes.get(); + + w.reset(); + std::vector be_results; + std::vector le_results; + std::vector ne_results; + be_results.reserve(numElements); + le_results.reserve(numElements); + ne_results.reserve(numElements); + for (int i = 0; i < numElements; ++i) + { + be_results.push_back(boost::int128::to_be_bytes(input_vector[i])); + le_results.push_back(boost::int128::to_le_bytes(input_vector[i])); + ne_results.push_back(boost::int128::to_ne_bytes(input_vector[i])); + } + double t = w.elapsed(); + + for (int i = 0; i < numElements; ++i) + { + if (output_be[i] != be_results[i] || output_le[i] != le_results[i] || output_ne[i] != ne_results[i]) + { + std::cerr << "Result verification failed at element " << i << "!" << std::endl; + return EXIT_FAILURE; + } + + const std::uint8_t *bytes = host_bytes + i * static_cast(sizeof(test_type)); + + if (output_from_be[i] != boost::int128::from_be_bytes(bytes) || + output_from_le[i] != boost::int128::from_le_bytes(bytes) || + output_from_ne[i] != boost::int128::from_ne_bytes(bytes)) + { + std::cerr << "Result verification failed at element " << i << "!" << std::endl; + return EXIT_FAILURE; + } + + // Reading back what was written recovers the value. The pointer overload is used + // rather than the std::array one because nvcc instantiates the device side of a + // __host__ __device__ template even for a host call, and the std::array overload + // reads its argument through std::array::operator[], a constexpr host function. + if (boost::int128::from_be_bytes(output_be[i].data()) != input_vector[i] || + boost::int128::from_le_bytes(output_le[i].data()) != input_vector[i] || + boost::int128::from_ne_bytes(output_ne[i].data()) != input_vector[i]) + { + std::cerr << "Round trip verification failed at element " << i << "!" << std::endl; + return EXIT_FAILURE; + } + } + + std::cout << "Test PASSED, normal calculation time: " << t << "s" << std::endl; + std::cout << "Done\n"; + + return 0; +} diff --git a/test/test_unsigned_byte_arrays_sycl.cpp b/test/test_unsigned_byte_arrays_sycl.cpp new file mode 100644 index 00000000..396b6081 --- /dev/null +++ b/test/test_unsigned_byte_arrays_sycl.cpp @@ -0,0 +1,12 @@ +// Copyright 2026 Matt Borland +// Distributed under the Boost Software License, Version 1.0. +// https://www.boost.org/LICENSE_1_0.txt + +#include "sycl_test.hpp" + +using boost::int128::uint128; + +int main() +{ + return int128_sycl_test::run_byte_arrays(); +} diff --git a/test/test_unsigned_byte_order.cu b/test/test_unsigned_byte_order.cu new file mode 100644 index 00000000..342b0d49 --- /dev/null +++ b/test/test_unsigned_byte_order.cu @@ -0,0 +1,98 @@ +// Copyright Matt Borland 2026. +// Use, modification and distribution are subject to the +// Boost Software License, Version 1.0. (See accompanying file +// LICENSE_1_0.txt or copy at http://www.boost.org/LICENSE_1_0.txt) + +#include +#include +#include +#include +#include +#include +#include +#include +#include "cuda_managed_ptr.hpp" +#include "stopwatch.hpp" + +#include + +using test_type = boost::int128::uint128; + +__global__ void cuda_test(const test_type *in, test_type *out_be, test_type *out_le, test_type *out_round, int numElements) +{ + int i = blockDim.x * blockIdx.x + threadIdx.x; + + if (i < numElements) + { + out_be[i] = boost::int128::to_be(in[i]); + out_le[i] = boost::int128::to_le(in[i]); + + // Both directions on the device, which must recover the input + out_round[i] = boost::int128::from_be(boost::int128::to_be(in[i])); + } +} + +int main(void) +{ + std::mt19937_64 rng {42}; + + cudaError_t err = cudaSuccess; + + int numElements = 50000; + std::cout << "[Vector operation on " << numElements << " elements]" << std::endl; + + cuda_managed_ptr input_vector(numElements); + cuda_managed_ptr output_be(numElements); + cuda_managed_ptr output_le(numElements); + cuda_managed_ptr output_round(numElements); + + boost::random::uniform_int_distribution dist {test_type{0U}, (std::numeric_limits::max)()}; + for (std::size_t i = 0; i < numElements; ++i) + { + input_vector[i] = dist(rng); + } + + int threadsPerBlock = 256; + int blocksPerGrid = (numElements + threadsPerBlock - 1) / threadsPerBlock; + std::cout << "CUDA kernel launch with " << blocksPerGrid << " blocks of " << threadsPerBlock << " threads" << std::endl; + + watch w; + + cuda_test<<>>(input_vector.get(), output_be.get(), output_le.get(), output_round.get(), numElements); + cudaDeviceSynchronize(); + + std::cout << "CUDA kernal done in: " << w.elapsed() << "s" << std::endl; + + err = cudaGetLastError(); + if (err != cudaSuccess) + { + std::cerr << "Failed to launch kernel (error code " << cudaGetErrorString(err) << ")!" << std::endl; + return EXIT_FAILURE; + } + + std::vector be_results; + std::vector le_results; + be_results.reserve(numElements); + le_results.reserve(numElements); + w.reset(); + for (int i = 0; i < numElements; ++i) + { + be_results.push_back(boost::int128::to_be(input_vector[i])); + le_results.push_back(boost::int128::to_le(input_vector[i])); + } + double t = w.elapsed(); + + for (int i = 0; i < numElements; ++i) + { + if (output_be[i] != be_results[i] || output_le[i] != le_results[i] || output_round[i] != input_vector[i]) + { + std::cerr << "Result verification failed at element " << i << "!" << std::endl; + return EXIT_FAILURE; + } + } + + std::cout << "Test PASSED, normal calculation time: " << t << "s" << std::endl; + std::cout << "Done\n"; + + return 0; +} diff --git a/test/test_unsigned_byte_order_sycl.cpp b/test/test_unsigned_byte_order_sycl.cpp new file mode 100644 index 00000000..a953389d --- /dev/null +++ b/test/test_unsigned_byte_order_sycl.cpp @@ -0,0 +1,12 @@ +// Copyright 2026 Matt Borland +// Distributed under the Boost Software License, Version 1.0. +// https://www.boost.org/LICENSE_1_0.txt + +#include "sycl_test.hpp" + +using boost::int128::uint128; + +int main() +{ + return int128_sycl_test::run_byte_order(); +} From 6ac38457c7318ebe2acd446ce069a52929e80ae3 Mon Sep 17 00:00:00 2001 From: Matt Borland Date: Mon, 3 Aug 2026 11:35:23 -0400 Subject: [PATCH 3/5] Add testing with the module --- module/CMakeLists.txt | 1 + module/Jamfile | 1 + module/int128.cppm | 1 + 3 files changed, 3 insertions(+) diff --git a/module/CMakeLists.txt b/module/CMakeLists.txt index 9715b540..49d5ecc1 100644 --- a/module/CMakeLists.txt +++ b/module/CMakeLists.txt @@ -61,6 +61,7 @@ message(STATUS "Boost.Int128 module: using 'import std;'") # support, so those two remain covered by the header-based test suite. set(BOOST_INT128_MODULE_TESTS test_bit + test_byte_conversions test_div test_from_chars_bases test_gcd_lcm diff --git a/module/Jamfile b/module/Jamfile index f3375450..cd5fe936 100644 --- a/module/Jamfile +++ b/module/Jamfile @@ -33,6 +33,7 @@ obj int128 : int128.cppm : BOOST_INT128_EXPORT_TESTING msvc:int128 ; run ../test/test_bit.cpp int128 : : : int128 ; +run ../test/test_byte_conversions.cpp int128 : : : int128 ; run ../test/test_div.cpp int128 : : : int128 ; run ../test/test_from_chars_bases.cpp int128 : : : int128 ; run ../test/test_gcd_lcm.cpp int128 : : : int128 ; diff --git a/module/int128.cppm b/module/int128.cppm index de6b35c9..c61ca1f2 100644 --- a/module/int128.cppm +++ b/module/int128.cppm @@ -36,6 +36,7 @@ module; // otherwise they are supplied here in the global module fragment. #ifndef BOOST_INT128_USE_STD_MODULE +#include #include #include #include From e7bd7fbdf987dc320f3348435b50994a388c55ad Mon Sep 17 00:00:00 2001 From: Matt Borland Date: Mon, 3 Aug 2026 11:35:28 -0400 Subject: [PATCH 4/5] Add an example --- examples/byte_conversions.cpp | 88 +++++++++++++++++++++++++++++++++++ 1 file changed, 88 insertions(+) create mode 100644 examples/byte_conversions.cpp diff --git a/examples/byte_conversions.cpp b/examples/byte_conversions.cpp new file mode 100644 index 00000000..014591b7 --- /dev/null +++ b/examples/byte_conversions.cpp @@ -0,0 +1,88 @@ +// Copyright 2026 Matt Borland +// Distributed under the Boost Software License, Version 1.0. +// https://www.boost.org/LICENSE_1_0.txt + +#include +#include +#include +#include +#include +#include +#include +#include + +// Prints the bytes of an array in the order they are stored +template +void print_bytes(const char* label, const Bytes& bytes) +{ + std::cout << label; + + for (const auto byte : bytes) + { + std::cout << ' ' << std::hex << std::setfill('0') << std::setw(2) << static_cast(byte); + } + + std::cout << std::dec << std::endl; +} + +int main() +{ + using boost::int128::uint128; + using boost::int128::int128; + + // The 16 bytes 01 02 ... 10 read as a big-endian value + constexpr uint128 value {UINT64_C(0x0102030405060708), UINT64_C(0x090A0B0C0D0E0F10)}; + + std::cout << "=== Byte arrays ===" << std::endl; + + // The byte order of the array is the requested one on every platform + print_bytes("to_be_bytes:", boost::int128::to_be_bytes(value)); + print_bytes("to_le_bytes:", boost::int128::to_le_bytes(value)); + + // Native order is whichever of the two matches the host, so this is the + // one form whose output depends on the platform + print_bytes("to_ne_bytes:", boost::int128::to_ne_bytes(value)); + + std::cout << "\n=== Reading a value back out of bytes ===" << std::endl; + + constexpr std::array wire + {{ + 0x00, 0x00, 0x00, 0x00, 0x00, 0x00, 0x00, 0x00, + 0x00, 0x00, 0x00, 0x00, 0x00, 0x00, 0x01, 0x2C + }}; + + // The target type is given explicitly, and the byte count has to match + std::cout << "from_be_bytes: " << boost::int128::from_be_bytes(wire) << std::endl; + std::cout << "from_le_bytes: " << boost::int128::from_le_bytes(wire) << std::endl; + + // Everything is constexpr, so a whole round-trip can be checked at compile time + static_assert(boost::int128::from_be_bytes(boost::int128::to_be_bytes(value)) == value, + "Round trip through big-endian bytes"); + + std::cout << "\n=== Signed values ===" << std::endl; + + // The two's complement bit pattern is what gets reversed, so negative + // values need no special handling + constexpr int128 negative {-300}; + + print_bytes("to_be_bytes(-300):", boost::int128::to_be_bytes(negative)); + std::cout << "from_be_bytes: " << boost::int128::from_be_bytes(boost::int128::to_be_bytes(negative)) << std::endl; + + std::cout << "\n=== Whole value conversions ===" << std::endl; + + // to_be and to_le produce a value whose object representation is in the + // requested order, which is what a memcpy into a packet buffer wants. + // The value itself is only meaningful again after the matching from_be / from_le. + const auto big_endian_image {boost::int128::to_be(value)}; + + print_bytes("object representation of to_be(value):", boost::int128::to_ne_bytes(big_endian_image)); + std::cout << "from_be recovers: " << boost::int128::from_be(big_endian_image) << std::endl; + std::cout << "value: " << value << std::endl; + + // Any byte-like element type can be requested, which is convenient when the + // surrounding buffer is not made of std::uint8_t + const auto as_char {boost::int128::to_le_bytes(value)}; + std::cout << "\nfrom_le_bytes over a char buffer: " << boost::int128::from_le_bytes(as_char.data()) << std::endl; + + return 0; +} From 043c74b1af7e24f1dce773f0b6fc7ac340c34873 Mon Sep 17 00:00:00 2001 From: Matt Borland Date: Mon, 3 Aug 2026 11:35:37 -0400 Subject: [PATCH 5/5] Update documentation --- doc/modules/ROOT/nav.adoc | 5 + doc/modules/ROOT/pages/api_reference.adoc | 41 +++ doc/modules/ROOT/pages/byte_conversions.adoc | 277 +++++++++++++++++++ doc/modules/ROOT/pages/examples.adoc | 15 + doc/modules/ROOT/pages/file_structure.adoc | 3 + doc/modules/ROOT/pages/release_notes.adoc | 19 ++ 6 files changed, 360 insertions(+) create mode 100644 doc/modules/ROOT/pages/byte_conversions.adoc diff --git a/doc/modules/ROOT/nav.adoc b/doc/modules/ROOT/nav.adoc index 38932363..0724e851 100644 --- a/doc/modules/ROOT/nav.adoc +++ b/doc/modules/ROOT/nav.adoc @@ -6,6 +6,7 @@ ** xref:examples.adoc#examples_io[IO Streaming] ** xref:examples.adoc#examples_rollover[Rollover Behavior] ** xref:examples.adoc#examples_bit[`` support (Bitwise Operations)] +** xref:examples.adoc#examples_byte_conversions[Byte Order Conversions] ** xref:examples.adoc#examples_numeric[`` support (Saturating Arithmetic)] ** xref:examples.adoc#examples_numeric_algorithms[`` support (Numeric Algorithms)] ** xref:examples.adoc#examples_integer_division[`` support (Integer Division)] @@ -34,6 +35,7 @@ ** xref:api_reference.adoc#api_structs[Structs] ** xref:api_reference.adoc#api_functions[Functions] *** xref:api_reference.adoc#api_bit[`` (Bitwise ops)] +*** xref:api_reference.adoc#api_byte_conversions[Byte Order Conversions] *** xref:api_reference.adoc#api_charconv[`` (`from_chars` and `to_chars`)] *** xref:api_reference.adoc#api_cmath[`` (`abs`)] *** xref:api_reference.adoc#api_cstdlib[`` (div and mod functions)] @@ -72,6 +74,9 @@ ** xref:mixed_type_ops.adoc#mixed_ops_floating_point[Operations With Floating-Point Types] * xref:literals.adoc[] * xref:bit.adoc[`` (Bitwise ops)] +* xref:byte_conversions.adoc[Byte Order Conversions] +** xref:byte_conversions.adoc#byte_conversions_element_type[Byte Array Element Type] +** xref:byte_conversions.adoc#byte_conversions_examples[Examples] * xref:cstdlib.adoc[`` (div and mod functions)] * xref:charconv.adoc[`` (`from_chars` and `to_chars`)] * xref:stream.adoc[`` support] diff --git a/doc/modules/ROOT/pages/api_reference.adoc b/doc/modules/ROOT/pages/api_reference.adoc index 9bd7b5dd..3acc3f29 100644 --- a/doc/modules/ROOT/pages/api_reference.adoc +++ b/doc/modules/ROOT/pages/api_reference.adoc @@ -138,6 +138,44 @@ Listed by analogous STL header. | Reverses byte order |=== +[#api_byte_conversions] +=== xref:byte_conversions.adoc[Byte Order Conversions] + +[cols="1,2", options="header"] +|=== +| Function | Description + +| xref:byte_conversions.adoc#to_be[`to_be`] +| Converts a value from native to big-endian byte order + +| xref:byte_conversions.adoc#from_be[`from_be`] +| Converts a value from big-endian to native byte order + +| xref:byte_conversions.adoc#to_le[`to_le`] +| Converts a value from native to little-endian byte order + +| xref:byte_conversions.adoc#from_le[`from_le`] +| Converts a value from little-endian to native byte order + +| xref:byte_conversions.adoc#to_be_bytes[`to_be_bytes`] +| Returns the bytes of a value, most significant first + +| xref:byte_conversions.adoc#from_be_bytes[`from_be_bytes`] +| Reconstructs a value from big-endian ordered bytes + +| xref:byte_conversions.adoc#to_le_bytes[`to_le_bytes`] +| Returns the bytes of a value, least significant first + +| xref:byte_conversions.adoc#from_le_bytes[`from_le_bytes`] +| Reconstructs a value from little-endian ordered bytes + +| xref:byte_conversions.adoc#to_ne_bytes[`to_ne_bytes`] +| Returns the bytes of a value in native order + +| xref:byte_conversions.adoc#from_ne_bytes[`from_ne_bytes`] +| Reconstructs a value from native ordered bytes +|=== + [#api_charconv] === xref:charconv.adoc[``] @@ -472,6 +510,9 @@ Listed by analogous STL header. | xref:bit.adoc[``] | Bit manipulation functions +| xref:byte_conversions.adoc[``] +| Big-endian and little-endian byte order conversions + | xref:charconv.adoc[``] | Character conversion functions diff --git a/doc/modules/ROOT/pages/byte_conversions.adoc b/doc/modules/ROOT/pages/byte_conversions.adoc new file mode 100644 index 00000000..859812eb --- /dev/null +++ b/doc/modules/ROOT/pages/byte_conversions.adoc @@ -0,0 +1,277 @@ +//// +Copyright 2026 Matt Borland +Distributed under the Boost Software License, Version 1.0. +https://www.boost.org/LICENSE_1_0.txt +//// + +[#byte_conversions] += Byte Order Conversions +:idprefix: byte_conversions_ + +[source,c++] +---- +#include +---- + +The library provides functions for converting `uint128` and `int128` to and from big-endian or little-endian byte order, which is what serializing a 128-bit value into a file format, a network packet, or a database column requires. +Unlike the functions in xref:bit.adoc[``], every one of these has both an unsigned and a signed overload: reversing a two's complement bit pattern is the same operation for either sign. + +All of these functions are `constexpr`, and are available using pass:[C++14] like the rest of the library. + +There are two families: + +* `to_be` / `from_be` / `to_le` / `from_le` operate on whole values. On the platform whose byte order is being asked for they are the identity, and on the other they are a xref:bit.adoc#byteswap[`byteswap`]. The result of `to_be` and `to_le` is a value whose *object representation* is in the requested order, so it is meant to be written out (or `memcpy`'d) rather than read as a number. +* `to_*_bytes` / `from_*_bytes` operate on 16-byte arrays. These never depend on the host byte order (with the sole exception of the `ne` pair), so `to_be_bytes` returns the most significant byte first on every platform. + +The byte arrays are built with shifts rather than from the object representation, so the byte-array functions give identical results on little-endian and big-endian hosts. + +[#byte_conversions_element_type] +== Byte Array Element Type + +The array functions are templated on the element type, which defaults to `std::uint8_t`. +The accepted types are `char`, `signed char`, `unsigned char`, and, when the standard library provides it (pass:[C++17]), `std::byte`: + +[source,c++] +---- +const auto bytes {boost::int128::to_be_bytes(value)}; // std::array +const auto as_byte {boost::int128::to_be_bytes(value)}; // std::array +const auto as_char {boost::int128::to_be_bytes(value)}; // std::array +---- + +NOTE: The number of bytes is fixed at `sizeof(T)`, which is 16, so the `from_*_bytes` functions do not need a length argument. The `std::array` overloads reject a mismatched size at compile time, and the pointer overloads read exactly 16 bytes from the address given. A `std::span`, `std::vector`, or any other contiguous range is passed by handing `.data()` to the pointer overload. + +NOTE: In a CUDA translation unit the pointer overloads of `from_be_bytes`, `from_le_bytes`, and `from_ne_bytes` are the portable choice. The `std::array` overloads read the array through `std::array::operator[]`, which is a `constexpr` host function, so `nvcc` warns (`#20013-D`) unless it is given `--expt-relaxed-constexpr`. The warning appears even when the call itself is host code, because `nvcc` compiles the device side of a `__host__ __device__` template as well. Writing bytes with `to_be_bytes`, `to_le_bytes`, or `to_ne_bytes` is unaffected: the returned array is constructed and assigned without calling any of its member functions. + +[#to_be] +== to_be + +Converts a value from the native byte order to big-endian byte order. +On a big-endian platform the value is returned unchanged, and on a little-endian platform `byteswap` is applied. + +[source,c++] +---- +namespace boost { +namespace int128 { + +BOOST_INT128_HOST_DEVICE constexpr uint128 to_be(uint128 value) noexcept; +BOOST_INT128_HOST_DEVICE constexpr int128 to_be(int128 value) noexcept; + +} // namespace int128 +} // namespace boost +---- + +[#from_be] +== from_be + +Converts a value from big-endian byte order to the native byte order. +This is the inverse of `to_be`, and since a byte reversal is its own inverse it delegates directly to `to_be`. + +[source,c++] +---- +namespace boost { +namespace int128 { + +BOOST_INT128_HOST_DEVICE constexpr uint128 from_be(uint128 value) noexcept; +BOOST_INT128_HOST_DEVICE constexpr int128 from_be(int128 value) noexcept; + +} // namespace int128 +} // namespace boost +---- + +[#to_le] +== to_le + +Converts a value from the native byte order to little-endian byte order. +On a little-endian platform the value is returned unchanged, and on a big-endian platform `byteswap` is applied. + +[source,c++] +---- +namespace boost { +namespace int128 { + +BOOST_INT128_HOST_DEVICE constexpr uint128 to_le(uint128 value) noexcept; +BOOST_INT128_HOST_DEVICE constexpr int128 to_le(int128 value) noexcept; + +} // namespace int128 +} // namespace boost +---- + +[#from_le] +== from_le + +Converts a value from little-endian byte order to the native byte order. +This is the inverse of `to_le`, and delegates directly to it. + +[source,c++] +---- +namespace boost { +namespace int128 { + +BOOST_INT128_HOST_DEVICE constexpr uint128 from_le(uint128 value) noexcept; +BOOST_INT128_HOST_DEVICE constexpr int128 from_le(int128 value) noexcept; + +} // namespace int128 +} // namespace boost +---- + +[#to_be_bytes] +== to_be_bytes + +Returns the 16 bytes of the value with the most significant byte first, on every platform. + +[source,c++] +---- +namespace boost { +namespace int128 { + +template +BOOST_INT128_HOST_DEVICE constexpr std::array to_be_bytes(uint128 value) noexcept; + +template +BOOST_INT128_HOST_DEVICE constexpr std::array to_be_bytes(int128 value) noexcept; + +} // namespace int128 +} // namespace boost +---- + +[#from_be_bytes] +== from_be_bytes + +Reconstructs a value from 16 bytes in big-endian order. +The target type is given explicitly and must be `uint128` or `int128`. + +[source,c++] +---- +namespace boost { +namespace int128 { + +template +BOOST_INT128_HOST_DEVICE constexpr T from_be_bytes(const std::array& bytes) noexcept; + +template +BOOST_INT128_HOST_DEVICE constexpr T from_be_bytes(const ByteType* bytes) noexcept; + +} // namespace int128 +} // namespace boost +---- + +The `std::array` overload requires `N == sizeof(T)`, and any other size is a `static_assert` failure. +The pointer overload reads `sizeof(T)` bytes starting at `bytes`, and the caller is responsible for that many bytes being readable. + +[#to_le_bytes] +== to_le_bytes + +Returns the 16 bytes of the value with the least significant byte first, on every platform. + +[source,c++] +---- +namespace boost { +namespace int128 { + +template +BOOST_INT128_HOST_DEVICE constexpr std::array to_le_bytes(uint128 value) noexcept; + +template +BOOST_INT128_HOST_DEVICE constexpr std::array to_le_bytes(int128 value) noexcept; + +} // namespace int128 +} // namespace boost +---- + +[#from_le_bytes] +== from_le_bytes + +Reconstructs a value from 16 bytes in little-endian order. +The size requirements match xref:byte_conversions.adoc#from_be_bytes[`from_be_bytes`]. + +[source,c++] +---- +namespace boost { +namespace int128 { + +template +BOOST_INT128_HOST_DEVICE constexpr T from_le_bytes(const std::array& bytes) noexcept; + +template +BOOST_INT128_HOST_DEVICE constexpr T from_le_bytes(const ByteType* bytes) noexcept; + +} // namespace int128 +} // namespace boost +---- + +[#to_ne_bytes] +== to_ne_bytes + +Returns the 16 bytes of the value in the native byte order, which is the object representation of the value. +Delegates to `to_le_bytes` on a little-endian platform and to `to_be_bytes` on a big-endian one, so this is the only byte-array function whose result varies across platforms. + +[source,c++] +---- +namespace boost { +namespace int128 { + +template +BOOST_INT128_HOST_DEVICE constexpr std::array to_ne_bytes(uint128 value) noexcept; + +template +BOOST_INT128_HOST_DEVICE constexpr std::array to_ne_bytes(int128 value) noexcept; + +} // namespace int128 +} // namespace boost +---- + +[#from_ne_bytes] +== from_ne_bytes + +Reconstructs a value from 16 bytes in the native byte order. +Delegates to `from_le_bytes` on a little-endian platform and to `from_be_bytes` on a big-endian one. +The size requirements match xref:byte_conversions.adoc#from_be_bytes[`from_be_bytes`]. + +[source,c++] +---- +namespace boost { +namespace int128 { + +template +BOOST_INT128_HOST_DEVICE constexpr T from_ne_bytes(const std::array& bytes) noexcept; + +template +BOOST_INT128_HOST_DEVICE constexpr T from_ne_bytes(const ByteType* bytes) noexcept; + +} // namespace int128 +} // namespace boost +---- + +[#byte_conversions_examples] +== Examples + +.This https://github.com/cppalliance/int128/blob/develop/examples/byte_conversions.cpp[example] demonstrates the byte order conversion functions +==== +[source, c++] +---- +include::example$byte_conversions.cpp[] +---- + +Output (on a little-endian host, where the `to_ne_bytes` lines match `to_le_bytes`): +---- +=== Byte arrays === +to_be_bytes: 01 02 03 04 05 06 07 08 09 0a 0b 0c 0d 0e 0f 10 +to_le_bytes: 10 0f 0e 0d 0c 0b 0a 09 08 07 06 05 04 03 02 01 +to_ne_bytes: 10 0f 0e 0d 0c 0b 0a 09 08 07 06 05 04 03 02 01 + +=== Reading a value back out of bytes === +from_be_bytes: 300 +from_le_bytes: 58491224111394833235396041148664381440 + +=== Signed values === +to_be_bytes(-300): ff ff ff ff ff ff ff ff ff ff ff ff ff ff fe d4 +from_be_bytes: -300 + +=== Whole value conversions === +object representation of to_be(value): 01 02 03 04 05 06 07 08 09 0a 0b 0c 0d 0e 0f 10 +from_be recovers: 1339673755198158349044581307228491536 +value: 1339673755198158349044581307228491536 + +from_le_bytes over a char buffer: 1339673755198158349044581307228491536 +---- +==== diff --git a/doc/modules/ROOT/pages/examples.adoc b/doc/modules/ROOT/pages/examples.adoc index 996fa23b..3c6b5b24 100644 --- a/doc/modules/ROOT/pages/examples.adoc +++ b/doc/modules/ROOT/pages/examples.adoc @@ -218,6 +218,21 @@ include::example$bit.cpp[tags=**;!exclude] ---- ==== +[#examples_byte_conversions] +== Byte Order Conversions + +Serializing a 128-bit value into a packet or a file format means pinning down its byte order. +This example covers both families: the whole value conversions (`to_be`, `from_be`, `to_le`, `from_le`), which are the identity on the matching platform and a byteswap on the other, and the byte array conversions (`to_be_bytes`, `from_be_bytes`, and their little and native endian counterparts), which give the same answer on every platform. +It also shows a signed value, a compile-time round-trip, and how to select a byte array element type other than `std::uint8_t`. + +.This https://github.com/cppalliance/int128/blob/develop/examples/byte_conversions.cpp[example] demonstrates the byte order conversion functions +==== +[source, c++] +---- +include::example$byte_conversions.cpp[] +---- +==== + [#examples_numeric] == Saturating Arithmetic () diff --git a/doc/modules/ROOT/pages/file_structure.adoc b/doc/modules/ROOT/pages/file_structure.adoc index 949b21df..e3907732 100644 --- a/doc/modules/ROOT/pages/file_structure.adoc +++ b/doc/modules/ROOT/pages/file_structure.adoc @@ -21,6 +21,9 @@ The entire library can be consumed via ``, or by independently | xref:bit.adoc[``] | Bit manipulation functions +| xref:byte_conversions.adoc[``] +| Big-endian and little-endian byte order conversions + | xref:charconv.adoc[``] | Character conversion (`to_chars`/`from_chars`); requires Boost.Charconv headers diff --git a/doc/modules/ROOT/pages/release_notes.adoc b/doc/modules/ROOT/pages/release_notes.adoc index 31e591df..9b067688 100644 --- a/doc/modules/ROOT/pages/release_notes.adoc +++ b/doc/modules/ROOT/pages/release_notes.adoc @@ -10,6 +10,25 @@ https://www.boost.org/LICENSE_1_0.txt == Unreleased +=== New: byte order conversions + +The new header `` converts `uint128` and `int128` to and from big-endian or little-endian byte order, which is what writing a 128-bit value into a packet, a file format, or a database column requires. + +`to_be` and `to_le` return a value whose object representation is in the requested order, and `from_be` and `from_le` read one back. +`to_be_bytes`, `to_le_bytes`, and `to_ne_bytes` return the 16 bytes as a `std::array`, and `from_be_bytes`, `from_le_bytes`, and `from_ne_bytes` reconstruct the value from either a `std::array` or a pointer to at least 16 bytes. + +[source, c++] +---- +constexpr uint128 value {UINT64_C(0x0102030405060708), UINT64_C(0x090A0B0C0D0E0F10)}; + +const auto bytes {to_be_bytes(value)}; // 01 02 03 ... 10 on every platform +static_assert(from_be_bytes(bytes) == value, "Round trip"); +---- + +Apart from the two native-endian functions, none of these depend on the host byte order, and both signs are supported since reversing a two's complement bit pattern is the same operation either way. +Every function is `constexpr` and available from pass:[C++14]. +See xref:byte_conversions.adoc[Byte Order Conversions]. + === Fix: the conversion to floating point is now correctly rounded on every platform Where the compiler provides no built-in 128-bit integer type, which is MSVC, the CUDA and SYCL device, and every 32-bit target, the conversion to `float` and `double` composed the value as `high * 2^64 + low`.