From 087c86b104149883c97ae274769c0dbc93638076 Mon Sep 17 00:00:00 2001 From: Matt Borland Date: Sun, 22 Mar 2026 13:17:31 -0500 Subject: [PATCH 1/3] Fix potential pessimization on host with CUDA detection --- include/boost/charconv/detail/config.hpp | 4 ++-- .../boost/charconv/detail/from_chars_integer_impl.hpp | 10 +++++----- include/boost/charconv/detail/memcpy.hpp | 2 +- .../boost/charconv/detail/to_chars_integer_impl.hpp | 8 ++++---- 4 files changed, 12 insertions(+), 12 deletions(-) diff --git a/include/boost/charconv/detail/config.hpp b/include/boost/charconv/detail/config.hpp index 59b71ba6..d4cba2cc 100644 --- a/include/boost/charconv/detail/config.hpp +++ b/include/boost/charconv/detail/config.hpp @@ -20,7 +20,7 @@ #endif // Use 128-bit integers and suppress warnings for using extensions -#if defined(BOOST_HAS_INT128) && !defined(__NVCC__) +#if defined(BOOST_HAS_INT128) && !(defined(BOOST_CHARCONV_ENABLE_CUDA) && defined(__CUDACC__)) # define BOOST_CHARCONV_HAS_INT128 # define BOOST_CHARCONV_INT128_MAX static_cast((static_cast(1) << 127) - 1) # define BOOST_CHARCONV_INT128_MIN (-BOOST_CHARCONV_INT128_MAX - 1) @@ -201,7 +201,7 @@ static_assert(std::is_same::value, "__float128 should b #endif -#ifdef __NVCC__ +#if defined(BOOST_CHARCONV_ENABLE_CUDA) && defined(__CUDACC__) # define BOOST_CHARCONV_HOST_DEVICE __host__ __device__ #else # define BOOST_CHARCONV_HOST_DEVICE diff --git a/include/boost/charconv/detail/from_chars_integer_impl.hpp b/include/boost/charconv/detail/from_chars_integer_impl.hpp index 4c95cf3f..6680c06a 100644 --- a/include/boost/charconv/detail/from_chars_integer_impl.hpp +++ b/include/boost/charconv/detail/from_chars_integer_impl.hpp @@ -22,7 +22,7 @@ namespace boost { namespace charconv { namespace detail { -#ifndef __NVCC__ +#if !(defined(BOOST_CHARCONV_ENABLE_CUDA) && defined(__CUDACC__)) static constexpr unsigned char uchar_values[] = {255, 255, 255, 255, 255, 255, 255, 255, 255, 255, 255, 255, 255, 255, 255, 255, @@ -90,7 +90,7 @@ static constexpr double log_2_table[] = // Convert characters for 0-9, A-Z, a-z to 0-35. Anything else is 255 BOOST_CHARCONV_HOST_DEVICE constexpr unsigned char digit_from_char(const char val) noexcept { - #ifdef __NVCC__ + #if defined(BOOST_CHARCONV_ENABLE_CUDA) && defined(__CUDACC__) constexpr unsigned char uchar_values[] = {255, 255, 255, 255, 255, 255, 255, 255, 255, 255, 255, 255, 255, 255, 255, 255, @@ -137,7 +137,7 @@ BOOST_CHARCONV_HOST_DEVICE constexpr unsigned char digit_from_char(const char va #endif -#ifdef __NVCC__ +#if defined(BOOST_CHARCONV_ENABLE_CUDA) && defined(__CUDACC__) template __host__ __device__ constexpr T get_max_value() @@ -161,7 +161,7 @@ constexpr T get_max_value() template BOOST_CHARCONV_HOST_DEVICE BOOST_CXX14_CONSTEXPR from_chars_result from_chars_integer_impl(const char* first, const char* last, Integer& value, int base) noexcept { - #ifdef __NVCC__ + #if defined(BOOST_CHARCONV_ENABLE_CUDA) && defined(__CUDACC__) constexpr double log_2_table[] = { @@ -417,7 +417,7 @@ BOOST_CHARCONV_GCC5_CONSTEXPR from_chars_result from_chars128(const char* first, } #endif -#ifndef __NVCC__ +#if !(defined(BOOST_CHARCONV_ENABLE_CUDA) && defined(__CUDACC__)) BOOST_CHARCONV_GCC5_CONSTEXPR from_chars_result from_chars128(const char* first, const char* last, uint128& value, int base = 10) noexcept { return from_chars_integer_impl(first, last, value, base); diff --git a/include/boost/charconv/detail/memcpy.hpp b/include/boost/charconv/detail/memcpy.hpp index 4f30dccd..e3edf587 100644 --- a/include/boost/charconv/detail/memcpy.hpp +++ b/include/boost/charconv/detail/memcpy.hpp @@ -25,7 +25,7 @@ namespace boost { namespace charconv { namespace detail { -#ifdef __NVCC__ +#if defined(BOOST_CHARCONV_ENABLE_CUDA) && defined(__CUDACC__) __host__ __device__ constexpr char* memcpy(char* dest, const char* src, std::size_t count) { diff --git a/include/boost/charconv/detail/to_chars_integer_impl.hpp b/include/boost/charconv/detail/to_chars_integer_impl.hpp index e6e7b3ca..79066d82 100644 --- a/include/boost/charconv/detail/to_chars_integer_impl.hpp +++ b/include/boost/charconv/detail/to_chars_integer_impl.hpp @@ -52,7 +52,7 @@ static constexpr char radix_table[] = { '9', '5', '9', '6', '9', '7', '9', '8', '9', '9' }; -#ifndef __NVCC__ +#if !(defined(BOOST_CHARCONV_ENABLE_CUDA) && defined(__CUDACC__)) static constexpr char digit_table[] = { '0', '1', '2', '3', '4', '5', '6', '7', '8', '9', @@ -319,7 +319,7 @@ BOOST_CHARCONV_CONSTEXPR to_chars_result to_chars_128integer_impl(char* first, c template BOOST_CHARCONV_HOST_DEVICE BOOST_CHARCONV_CONSTEXPR to_chars_result to_chars_integer_impl(char* first, char* last, Integer value, int base) noexcept { - #ifdef __NVCC__ + #if defined(BOOST_CHARCONV_ENABLE_CUDA) && defined(__CUDACC__) constexpr char digit_table[] = { '0', '1', '2', '3', '4', '5', '6', '7', '8', '9', @@ -396,7 +396,7 @@ BOOST_CHARCONV_HOST_DEVICE BOOST_CHARCONV_CONSTEXPR to_chars_result to_chars_int } break; - #ifdef __NVCC__ + #if defined(BOOST_CHARCONV_ENABLE_CUDA) && defined(__CUDACC__) case 10: while (unsigned_value != static_cast(0)) @@ -463,7 +463,7 @@ BOOST_CHARCONV_HOST_DEVICE BOOST_CHARCONV_CONSTEXPR to_chars_result to_chars_int // The specialized base 10 path requires lookup tables and memcpy // On device, we instead use the trivial divide and mod to avoid these - #ifndef __NVCC__ + #if !(defined(BOOST_CHARCONV_ENABLE_CUDA) && defined(__CUDACC__)) if (base == 10) { return to_chars_integer_impl(first, last, value); From 50789b06118627eb85630c733de2d0f7fb064c50 Mon Sep 17 00:00:00 2001 From: Matt Borland Date: Sun, 22 Mar 2026 13:18:46 -0500 Subject: [PATCH 2/3] Pass macro enabling CUDA as COMPILE_DEFINITION --- test/CMakeLists.txt | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/test/CMakeLists.txt b/test/CMakeLists.txt index 1b6afc06..1bd497dd 100644 --- a/test/CMakeLists.txt +++ b/test/CMakeLists.txt @@ -16,7 +16,7 @@ if(HAVE_BOOST_TEST) enable_language(CUDA) set(CMAKE_CUDA_EXTENSIONS OFF) - boost_test_jamfile(FILE cuda_jamfile LINK_LIBRARIES Boost::charconv Boost::core Boost::assert ${CUDA_LIBRARIES} INCLUDE_DIRECTORIES ${CUDA_INCLUDE_DIRS} ) + boost_test_jamfile(FILE cuda_jamfile LINK_LIBRARIES Boost::charconv Boost::core Boost::assert ${CUDA_LIBRARIES} COMPILE_DEFINITIONS BOOST_CHARCONV_ENABLE_CUDA=1 INCLUDE_DIRECTORIES ${CUDA_INCLUDE_DIRS} ) else() From 68126c89dc77d7e8e98f03de64e6636ac1b4c674 Mon Sep 17 00:00:00 2001 From: Matt Borland Date: Sun, 22 Mar 2026 13:23:04 -0500 Subject: [PATCH 3/3] Better document CUDA support and how to use --- doc/charconv/api_reference.adoc | 2 ++ doc/charconv/build.adoc | 6 ++++++ 2 files changed, 8 insertions(+) diff --git a/doc/charconv/api_reference.adoc b/doc/charconv/api_reference.adoc index 39fe9c23..dfae7987 100644 --- a/doc/charconv/api_reference.adoc +++ b/doc/charconv/api_reference.adoc @@ -30,5 +30,7 @@ https://www.boost.org/LICENSE_1_0.txt == Macros +- <> +- <> - <> - <> diff --git a/doc/charconv/build.adoc b/doc/charconv/build.adoc index d2844309..d9de8986 100644 --- a/doc/charconv/build.adoc +++ b/doc/charconv/build.adoc @@ -36,6 +36,12 @@ If you are using another build system and you want support for these types you w IMPORTANT: libquadmath is only available on supported platforms (e.g. Linux with x86, x86_64, PPC64, and IA64). +[#enable_cuda_] +== CUDA Support + +This library has partial support for CUDA which can be enabled during compilation with `BOOST_CHARCONV_ENABLE_CUDA`. +Functions with `BOOST_CHARCONV_HOST_DEVICE` in their signature can be run on both host and device, all others are strictly run on host. + == Dependencies This library depends on: Boost.Assert, Boost.Config, Boost.Core, and optionally libquadmath (see above).