Skip to content

Fix CUDA pessimization and improve docs #4

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

Merged
mborland merged 3 commits into from
Mar 22, 2026
Merged

Fix CUDA pessimization and improve docs #4

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
3 commits
Select commit Hold shift + click to select a range
087c86b
Fix potential pessimization on host with CUDA detection
mborland Mar 22, 2026
50789b0
Pass macro enabling CUDA as COMPILE_DEFINITION
mborland Mar 22, 2026
68126c8
Better document CUDA support and how to use
mborland Mar 22, 2026
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
2 changes: 2 additions & 0 deletions doc/charconv/api_reference.adoc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -30,5 +30,7 @@ https://www.boost.org/LICENSE_1_0.txt

== Macros

- <<enable_cuda_, `BOOST_CHARCONV_ENABLE_CUDA`>>
- <<enable_cuda_, `BOOST_CHARCONV_HOST_DEVICE`>>
- <<integral_usage_notes_, `BOOST_CHARCONV_CONSTEXPR`>>
- <<run_benchmarks_, `BOOST_CHARCONV_RUN_BENCHMARKS`>>
6 changes: 6 additions & 0 deletions doc/charconv/build.adoc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -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).
4 changes: 2 additions & 2 deletions include/boost/charconv/detail/config.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -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<boost::int128_type>((static_cast<boost::uint128_type>(1) << 127) - 1)
# define BOOST_CHARCONV_INT128_MIN (-BOOST_CHARCONV_INT128_MAX - 1)
Expand Down Expand Up @@ -201,7 +201,7 @@ static_assert(std::is_same<long double, __float128>::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
Expand Down
10 changes: 5 additions & 5 deletions include/boost/charconv/detail/from_chars_integer_impl.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -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,
Expand Down Expand Up @@ -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,
Expand Down Expand Up @@ -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 <typename T>
__host__ __device__ constexpr T get_max_value()
Expand All @@ -161,7 +161,7 @@ constexpr T get_max_value()
template <typename Integer, typename Unsigned_Integer>
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[] =
{
Expand Down Expand Up @@ -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<uint128, uint128>(first, last, value, base);
Expand Down
2 changes: 1 addition & 1 deletion include/boost/charconv/detail/memcpy.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -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)
{
Expand Down
8 changes: 4 additions & 4 deletions include/boost/charconv/detail/to_chars_integer_impl.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -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',
Expand Down Expand Up @@ -319,7 +319,7 @@ BOOST_CHARCONV_CONSTEXPR to_chars_result to_chars_128integer_impl(char* first, c
template <typename Integer, typename Unsigned_Integer>
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',
Expand Down Expand Up @@ -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<Unsigned_Integer>(0))
Expand Down Expand Up @@ -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);
Expand Down
2 changes: 1 addition & 1 deletion test/CMakeLists.txt
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -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()

Expand Down
Loading