From 0af7022b5a81ac07d08a351fa29fd802688785f6 Mon Sep 17 00:00:00 2001 From: Allison Vacanti Date: Mon, 24 Feb 2020 18:57:47 -0500 Subject: [PATCH] Deprecate C++03, C++11, MSVC < 2017, GCC < 5.0 MIME-Version: 1.0 Content-Type: text/plain; charset=UTF-8 Content-Transfer-Encoding: 8bit Build infrastructure and static configuration fixes: - Bump CMAKE_CXX_STANDARD to 14 - Add `-Werror all-warnings` to NVCC to promote warnings to errors - Add `-Xcudafe --display_error_number` to get useful diagnositics from cudafe. - Clean up cub include dir spec in CMake. - Move THRUST_DEPRECATED logic out of compiler.h and into new header. - Fix CPP dialect detection on newer MSVC. - Remove raw `__cplusplus` checks. - Use `_Pragma`/`__pragma` instead of `#pragma` in macro. - Remove THRUST_BEGIN/END_NS macros. - These were used inconsistently, rendering them non-functional. Removing to prevent people from trying to use them. Workarounds for msvc: - MSVC isn't a fan of `decltype(...)::some_member` syntax. - WAR by aliasing the `decltype(...)` and doing `NewAlias::some_member` - Missing `template` keyword when rebinding pointer in `async/reduce.h` - Silence warning C4494 `declspec(allocator) used on non-pointer/ref type` - Bug in MSVC STL: https://github.com/microsoft/STL/issues/696 - Disable async sort test on MSVC - Triage. Looks like a bug in cudafe? See thrust/thrust#1098. - Add pointer::pointer_to(reference) - Required for C++11, hard compile error on MSVC. - Bring a definition of `atanh` into scope for complex number impl - Fix floating point literals be declared as floats instead of doubles - Replace `std::remove_reference::type&` with `std::add_lvalue_reference`. - Same behavior, and MSVC chokes on the other syntax when followed by `__host__`. - Remove constexpr markup from defaulted functions. - These are constexpr by default when possible, and the compilers were complaining about the markup in places. - Use `thrust::detail::integer_traits` instead of `std::numeric_limits` in device code. - Avoid aligning beyond platform limits in alignment.cu. - Pass /bigobj to MSVC so it can handle the async tests - Work around MSVC compiler bug by replacing SFINAE with static dispatch Bug 2865172 Bug 2880936 Reviewed-by: Bryce Adelstein Lelbach aka wash Reviewed-by: Michał 'Griwes' Dominiak --- CMakeLists.txt | 327 +++++++++++++++++- dependencies/cub | 2 +- doc/thrust.dox | 2 - examples/cuda/async_reduce.cu | 78 +++++ examples/cuda/global_device_vector.cu | 3 +- internal/benchmark/bench.cu | 4 +- internal/build/common_compiler.mk | 11 + testing/alignment.cu | 17 +- testing/allocator.cu | 5 +- testing/async_sort.cu | 4 +- testing/complex.cu | 4 +- testing/dependencies_aware_policies.cu | 7 +- testing/mr_disjoint_pool.cu | 10 +- testing/mr_pool.cu | 10 +- testing/vector.cu | 5 +- testing/vector_allocators.cu | 8 +- thrust/addressof.h | 5 +- thrust/allocate_unique.h | 5 +- thrust/async/copy.h | 5 +- thrust/async/for_each.h | 5 +- thrust/async/reduce.h | 92 +++-- thrust/async/sort.h | 28 +- thrust/async/transform.h | 5 +- thrust/complex.h | 2 +- thrust/detail/alignment.h | 12 +- thrust/detail/allocator/allocator_traits.h | 2 +- .../detail/allocator_aware_execution_policy.h | 2 +- thrust/detail/complex/catrig.h | 14 +- thrust/detail/complex/catrigf.h | 13 +- thrust/detail/config/compiler.h | 11 +- thrust/detail/config/config.h | 1 + thrust/detail/config/cpp_dialect.h | 104 +++++- .../{util/blocking.h => config/deprecated.h} | 34 +- thrust/detail/config/exec_check_disable.h | 6 +- thrust/detail/contiguous_storage.h | 5 +- thrust/detail/contiguous_storage.inl | 5 +- thrust/detail/distance.inl | 1 + thrust/detail/event_error.h | 5 +- thrust/detail/execute_with_allocator.h | 2 +- thrust/detail/execute_with_dependencies.h | 2 +- thrust/detail/memory_algorithms.h | 8 +- thrust/detail/pointer.h | 6 + thrust/detail/select_system.h | 5 +- thrust/detail/static_assert.h | 5 +- thrust/detail/tuple_algorithms.h | 5 +- thrust/detail/type_traits.h | 9 +- .../result_of_adaptable_function.h | 4 +- thrust/detail/vector_base.h | 5 +- thrust/detail/vector_base.inl | 5 +- thrust/device_make_unique.h | 5 +- thrust/future.h | 5 +- thrust/host_vector.h | 4 +- thrust/iterator/detail/reverse_iterator.inl | 4 +- thrust/iterator/reverse_iterator.h | 4 +- thrust/limits.h | 5 +- thrust/mr/allocator.h | 6 +- thrust/mr/detail/config.h | 2 +- thrust/mr/validator.h | 2 +- thrust/optional.h | 22 +- thrust/per_device_resource.h | 5 +- thrust/system/cpp/detail/vector.inl | 4 +- thrust/system/cpp/vector.h | 4 +- .../system/cuda/detail/adjacent_difference.h | 5 +- thrust/system/cuda/detail/assign_value.h | 5 +- thrust/system/cuda/detail/async/copy.h | 27 +- .../system/cuda/detail/async/customization.h | 5 +- thrust/system/cuda/detail/async/for_each.h | 5 +- thrust/system/cuda/detail/async/reduce.h | 7 +- thrust/system/cuda/detail/async/sort.h | 5 +- thrust/system/cuda/detail/async/transform.h | 5 +- thrust/system/cuda/detail/binary_search.h | 5 +- thrust/system/cuda/detail/copy.h | 10 +- thrust/system/cuda/detail/copy_if.h | 5 +- .../system/cuda/detail/core/agent_launcher.h | 5 +- thrust/system/cuda/detail/core/alignment.h | 5 +- .../cuda/detail/core/triple_chevron_launch.h | 5 +- thrust/system/cuda/detail/core/util.h | 5 +- thrust/system/cuda/detail/count.h | 5 +- thrust/system/cuda/detail/cross_system.h | 65 ++-- thrust/system/cuda/detail/dispatch.h | 7 +- thrust/system/cuda/detail/equal.h | 5 +- thrust/system/cuda/detail/execution_policy.h | 5 +- thrust/system/cuda/detail/extrema.h | 5 +- thrust/system/cuda/detail/fill.h | 5 +- thrust/system/cuda/detail/find.h | 10 +- thrust/system/cuda/detail/for_each.h | 5 +- thrust/system/cuda/detail/future.inl | 5 +- thrust/system/cuda/detail/gather.h | 5 +- thrust/system/cuda/detail/generate.h | 5 +- thrust/system/cuda/detail/get_value.h | 5 +- thrust/system/cuda/detail/inner_product.h | 5 +- .../cuda/detail/internal/copy_cross_system.h | 5 +- .../detail/internal/copy_device_to_device.h | 5 +- thrust/system/cuda/detail/iter_swap.h | 5 +- .../cuda/detail/make_unsigned_special.h | 5 +- thrust/system/cuda/detail/malloc_and_free.h | 5 +- thrust/system/cuda/detail/merge.h | 5 +- thrust/system/cuda/detail/mismatch.h | 10 +- thrust/system/cuda/detail/par.h | 5 +- thrust/system/cuda/detail/par_to_seq.h | 5 +- thrust/system/cuda/detail/parallel_for.h | 5 +- thrust/system/cuda/detail/partition.h | 5 +- .../system/cuda/detail/per_device_resource.h | 5 +- thrust/system/cuda/detail/reduce.h | 5 +- thrust/system/cuda/detail/reduce_by_key.h | 5 +- thrust/system/cuda/detail/remove.h | 5 +- thrust/system/cuda/detail/replace.h | 5 +- thrust/system/cuda/detail/reverse.h | 10 +- thrust/system/cuda/detail/scan.h | 10 +- thrust/system/cuda/detail/scan_by_key.h | 5 +- thrust/system/cuda/detail/scatter.h | 5 +- thrust/system/cuda/detail/set_operations.h | 5 +- thrust/system/cuda/detail/sort.h | 5 +- thrust/system/cuda/detail/swap_ranges.h | 5 +- thrust/system/cuda/detail/tabulate.h | 5 +- thrust/system/cuda/detail/transform.h | 5 +- thrust/system/cuda/detail/transform_reduce.h | 5 +- thrust/system/cuda/detail/transform_scan.h | 5 +- .../system/cuda/detail/uninitialized_copy.h | 5 +- .../system/cuda/detail/uninitialized_fill.h | 5 +- thrust/system/cuda/detail/unique.h | 5 +- thrust/system/cuda/detail/unique_by_key.h | 5 +- thrust/system/cuda/detail/util.h | 5 +- thrust/system/cuda/detail/vector.inl | 4 +- thrust/system/cuda/future.h | 5 +- thrust/system/cuda/memory.h | 5 +- thrust/system/cuda/memory_resource.h | 5 +- thrust/system/cuda/vector.h | 4 +- thrust/system/detail/generic/distance.inl | 2 +- thrust/system/omp/detail/vector.inl | 4 +- thrust/system/omp/vector.h | 4 +- thrust/system/tbb/detail/vector.inl | 4 +- thrust/system/tbb/vector.h | 4 +- thrust/type_traits/integer_sequence.h | 5 +- thrust/type_traits/is_contiguous_iterator.h | 11 +- thrust/type_traits/is_execution_policy.h | 5 +- ...operator_less_or_greater_function_object.h | 5 +- .../is_operator_plus_function_object.h | 5 +- thrust/type_traits/is_trivially_relocatable.h | 9 +- thrust/type_traits/logical_metafunctions.h | 5 +- thrust/type_traits/remove_cvref.h | 5 +- thrust/type_traits/void_t.h | 5 +- thrust/version.h | 12 - thrust/zip_function.h | 5 +- 144 files changed, 1095 insertions(+), 421 deletions(-) create mode 100644 examples/cuda/async_reduce.cu rename thrust/detail/{util/blocking.h => config/deprecated.h} (52%) diff --git a/CMakeLists.txt b/CMakeLists.txt index 2efccdb2a..29b29324f 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -62,13 +62,110 @@ if (NOT THRUST_HOST_SYSTEM IN_LIST THRUST_HOST_SYSTEM_OPTIONS) ) endif () -# Set CXX flags -set(CMAKE_CXX_STANDARD 11) -set(CMAKE_CXX_STANDARD_REQUIRED ON) +add_definitions(-DTHRUST_HOST_SYSTEM=THRUST_HOST_SYSTEM_${THRUST_HOST_SYSTEM}) + +set(THRUST_DEVICE_SYSTEM_OPTIONS CUDA CPP OMP TBB) +set(THRUST_DEVICE_SYSTEM CUDA CACHE STRING "The device backend to target.") +set_property( + CACHE THRUST_DEVICE_SYSTEM + PROPERTY STRINGS ${THRUST_DEVICE_SYSTEM_OPTIONS} +) +if (NOT THRUST_DEVICE_SYSTEM IN_LIST THRUST_DEVICE_SYSTEM_OPTIONS) + message( + FATAL_ERROR + "THRUST_DEVICE_SYSTEM must be one of ${THRUST_DEVICE_SYSTEM_OPTIONS}" + ) +endif () + +add_definitions(-DTHRUST_DEVICE_SYSTEM=THRUST_DEVICE_SYSTEM_${THRUST_DEVICE_SYSTEM}) + +# Please note this also sets the default for the CUDA C++ version; see the comment below. +set(CMAKE_CXX_STANDARD 14 CACHE STRING "The C++ version to be used.") set(CMAKE_CXX_EXTENSIONS OFF) -# Get dependencies -include(cmake/Dependencies.cmake) +message("-- C++ Standard version: ${CMAKE_CXX_STANDARD}") + +set(CUB_INCLUDE_DIR "${PROJECT_SOURCE_DIR}/dependencies/cub") + +if ("CUDA" STREQUAL "${THRUST_DEVICE_SYSTEM}") + if (NOT "${CMAKE_CUDA_HOST_COMPILER}" STREQUAL "") + unset(CMAKE_CUDA_HOST_COMPILER CACHE) + message(FATAL_ERROR "Thrust tests and examples require the C++ compiler" + " and the CUDA host compiler to be the same; to set this compiler, please" + " use the CMAKE_CXX_COMPILER variable, not the CMAKE_CUDA_HOST_COMPILER" + " variable.") + endif () + set(CMAKE_CUDA_HOST_COMPILER ${CMAKE_CXX_COMPILER}) + + enable_language(CUDA) + + # Force CUDA C++ standard to be the same as the C++ standard used. + # + # Now, CMake is unaligned with reality on standard versions: https://gitlab.kitware.com/cmake/cmake/issues/18597 + # which means that using standard CMake methods, it's impossible to actually sync the CXX and CUDA versions for pre-11 + # versions of C++; CUDA accepts 98 but translates that to 03, while CXX doesn't accept 03 (and doesn't translate that to 03). + # In case this gives You, dear user, any trouble, please escalate the above CMake bug, so we can support reality properly. + if (DEFINED CMAKE_CUDA_STANDARD) + message(WARNING "You've set CMAKE_CUDA_STANDARD; please note that this variable is ignored, and CMAKE_CXX_STANDARD" + " is used as the C++ standard version for both C++ and CUDA.") + endif() + unset(CMAKE_CUDA_STANDARD CACHE) + set(CMAKE_CUDA_STANDARD ${CMAKE_CXX_STANDARD}) + + set(THRUST_HIGHEST_COMPUTE_ARCH 75) + set(THRUST_KNOWN_COMPUTE_ARCHS 30 32 35 50 52 53 60 61 62 70 72 75) + + option(THRUST_DISABLE_ARCH_BY_DEFAULT "If ON, then all CUDA architectures are disabled on the initial CMake run." OFF) + set(OPTION_INIT ON) + if (THRUST_DISABLE_ARCH_BY_DEFAULT) + set(OPTION_INIT OFF) + endif () + + if (NOT ${THRUST_HIGHEST_COMPUTE_ARCH} IN_LIST THRUST_KNOWN_COMPUTE_ARCHS) + message(FATAL_ERROR "When changing the highest compute version, don't forget to add it to the list!") + endif () + + foreach (COMPUTE_ARCH IN LISTS THRUST_KNOWN_COMPUTE_ARCHS) + option(THRUST_ENABLE_COMPUTE_${COMPUTE_ARCH} "Enable code generation for tests for sm_${COMPUTE_ARCH}" ${OPTION_INIT}) + if (THRUST_ENABLE_COMPUTE_${COMPUTE_ARCH}) + set(CMAKE_CUDA_FLAGS "${CMAKE_CUDA_FLAGS} -gencode arch=compute_${COMPUTE_ARCH},code=sm_${COMPUTE_ARCH}") + set(COMPUTE_MESSAGE "${COMPUTE_MESSAGE} sm_${COMPUTE_ARCH}") + endif () + endforeach () + + option(THRUST_ENABLE_COMPUTE_FUTURE "Enable code generation for tests for compute_${THRUST_HIGHEST_COMPUTE_ARCH}" ${OPTION_INIT}) + if (THRUST_ENABLE_COMPUTE_FUTURE) + set(CMAKE_CUDA_FLAGS + "${CMAKE_CUDA_FLAGS} -gencode arch=compute_${THRUST_HIGHEST_COMPUTE_ARCH},code=compute_${THRUST_HIGHEST_COMPUTE_ARCH}") + set(COMPUTE_MESSAGE "${COMPUTE_MESSAGE} compute_${THRUST_HIGHEST_COMPUTE_ARCH}") + endif () + + message("-- Enabled CUDA architectures:${COMPUTE_MESSAGE}") +endif () + +if ("OMP" STREQUAL "${THRUST_DEVICE_SYSTEM}") + find_package(OpenMP REQUIRED) + if (OPENMP_FOUND) + set (CMAKE_C_FLAGS "${CMAKE_C_FLAGS} ${OpenMP_C_FLAGS}") + set (CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} ${OpenMP_CXX_FLAGS}") + set (CMAKE_EXE_LINKER_FLAGS "${CMAKE_EXE_LINKER_FLAGS} ${OpenMP_EXE_LINKER_FLAGS}") + endif() +endif () + +if ("TBB" STREQUAL "${THRUST_DEVICE_SYSTEM}") + find_package(PkgConfig REQUIRED) + pkg_check_modules(TBB tbb REQUIRED) + if (TBB_FOUND) + set (CMAKE_C_FLAGS "${CMAKE_C_FLAGS} ${TDD_CFLAGS}") + set (CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} ${TDD_CFLAGS}") + set (CMAKE_EXE_LINKER_FLAGS "${CMAKE_EXE_LINKER_FLAGS} ${TBB_LD_FLAGS}") + set (THRUST_ADDITIONAL_LIBRARIES "${TBB_LIBRARIES}") + endif () + + # There's a ton of these in the TBB backend, even though the code is correct. + # TODO: silence these warnings in code instead + append_option_if_available("-Wno-unused-parameter" THRUST_CXX_WARNINGS) +endif () # Setup VERSION rocm_setup_version(VERSION "2.10.6") @@ -98,8 +195,35 @@ if(BUILD_BENCHMARKS) add_subdirectory(internal/benchmark) endif() -set(THRUST_OPTIONS_DEBUG ${THRUST_OPTIONS_WARNINGS}) -set(THRUST_OPTIONS_RELEASE ${THRUST_OPTIONS_WARNINGS}) + # MSVC STL assumes that `allocator_traits`'s allocator will use raw pointers, + # and the `__DECLSPEC_ALLOCATOR` macro causes issues with thrust's universal + # allocators: + # warning C4494: 'std::allocator_traits<_Alloc>::allocate' : + # Ignoring __declspec(allocator) because the function return type is not + # a pointer or reference + # See https://github.com/microsoft/STL/issues/696 + append_option_if_available("/wd4494" THRUST_CXX_WARNINGS) + + # Some of the async tests require /bigobj to fit all their sections into the + # object files: + append_option_if_available("/bigobj" THRUST_CXX_WARNINGS) + + set(THRUST_TREAT_FILE_AS_CXX "/TP") +else () + append_option_if_available("-Werror" THRUST_CXX_WARNINGS) + append_option_if_available("-Wall" THRUST_CXX_WARNINGS) + append_option_if_available("-Wextra" THRUST_CXX_WARNINGS) + append_option_if_available("-Winit-self" THRUST_CXX_WARNINGS) + append_option_if_available("-Woverloaded-virtual" THRUST_CXX_WARNINGS) + append_option_if_available("-Wcast-qual" THRUST_CXX_WARNINGS) + append_option_if_available("-Wno-cast-align" THRUST_CXX_WARNINGS) + append_option_if_available("-Wno-long-long" THRUST_CXX_WARNINGS) + append_option_if_available("-Wno-variadic-macros" THRUST_CXX_WARNINGS) + append_option_if_available("-Wno-unused-function" THRUST_CXX_WARNINGS) + append_option_if_available("-Wno-unused-variable" THRUST_CXX_WARNINGS) + + set(THRUST_TREAT_FILE_AS_CXX "-x c++") +endif () # Package set(CPACK_DEBIAN_ARCHIVE_TYPE "gnutar") @@ -115,7 +239,186 @@ if(NOT CPACK_PACKAGING_INSTALL_PREFIX) set(CPACK_PACKAGING_INSTALL_PREFIX "${CMAKE_INSTALL_PREFIX}") endif() -set(CPACK_RPM_EXCLUDE_FROM_AUTO_FILELIST_ADDITION "\${CPACK_PACKAGING_INSTALL_PREFIX}" "\${CPACK_PACKAGING_INSTALL_PREFIX}/include") +if (("Clang" STREQUAL "${CMAKE_CXX_COMPILER_ID}") OR + ("XL" STREQUAL "${CMAKE_CXX_COMPILER_ID}")) + # xlC and Clang warn about unused parameters in uninstantiated templates. + # This causes xlC to choke on the OMP backend, which is mostly #ifdef'd out + # (and thus has unused parameters) when you aren't using it. + append_option_if_available("-Wno-unused-parameters" THRUST_CXX_WARNINGS) +endif () + +if ("Clang" STREQUAL "${CMAKE_CXX_COMPILER_ID}") + # -Wunneeded-internal-declaration misfires in the unit test framework + # on older versions of Clang. + append_option_if_available("-Wno-unneeded-internal-declaration" THRUST_CXX_WARNINGS) +endif () + +foreach (CXX_OPTION IN LISTS THRUST_CXX_WARNINGS) + set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} ${CXX_OPTION}") +endforeach () + +if ("CUDA" STREQUAL "${THRUST_DEVICE_SYSTEM}") + foreach (CXX_OPTION IN LISTS THRUST_CXX_WARNINGS) + set(CMAKE_CUDA_FLAGS "${CMAKE_CUDA_FLAGS} -Xcompiler=${CXX_OPTION}") + endforeach () + set(CMAKE_CUDA_FLAGS + "${CMAKE_CUDA_FLAGS} -Werror all-warnings -Xcudafe --display_error_number") +endif () + +# For every public header, build a translation unit containing `#include
` +# to let the compiler try to figure out warnings in that header if it is not otherwise +# included in tests, and also to verify if the headers are modular enough. +# .inl files are not globbed for, because they are not supposed to be used as public +# entrypoints. +list(APPEND THRUST_HEADER_GLOBS thrust/*.h) +list(APPEND THRUST_HEADER_EXCLUDE_SYSTEMS_GLOBS thrust/system/*/*) + +string(TOLOWER ${THRUST_HOST_SYSTEM} THRUST_HOST_SYSTEM_LOWERCASE) +list(APPEND THRUST_HEADER_SYSTEMS_GLOBS thrust/system/${THRUST_HOST_SYSTEM_LOWERCASE}/*) + +string(TOLOWER ${THRUST_DEVICE_SYSTEM} THRUST_DEVICE_SYSTEM_LOWERCASE) +list(APPEND THRUST_HEADER_SYSTEMS_GLOBS thrust/system/${THRUST_DEVICE_SYSTEM_LOWERCASE}/*) + +list(APPEND THRUST_HEADER_EXCLUDE_DETAILS_GLOBS thrust/detail/*) +list(APPEND THRUST_HEADER_EXCLUDE_DETAILS_GLOBS thrust/*/detail/*) +list(APPEND THRUST_HEADER_EXCLUDE_DETAILS_GLOBS thrust/*/*/detail/*) + +# Get all .h files... +file( + GLOB_RECURSE THRUST_HEADERS + RELATIVE ${PROJECT_SOURCE_DIR}/thrust + ${CMAKE_CONFIGURE_DEPENDS} + ${THRUST_HEADER_GLOBS} +) + +# ...then remove all system specific headers... +file( + GLOB_RECURSE THRUST_HEADER_EXCLUDE_SYSTEMS + RELATIVE ${PROJECT_SOURCE_DIR}/thrust + ${CMAKE_CONFIGURE_DEPENDS} + ${THRUST_HEADER_EXCLUDE_SYSTEMS_GLOBS} +) +list(REMOVE_ITEM THRUST_HEADERS ${THRUST_HEADER_EXCLUDE_SYSTEMS}) + +# ...then add all headers specific to the selected host and device systems back again... +file( + GLOB_RECURSE THRUST_SYSTEMS_HEADERS + RELATIVE ${PROJECT_SOURCE_DIR}/thrust + ${CMAKE_CONFIGURE_DEPENDS} + ${THRUST_HEADER_SYSTEMS_GLOBS} +) +list(APPEND THRUST_HEADERS ${THRUST_SYSTEMS_HEADERS}) + +# ...and remove all the detail headers (also removing the detail headers from the selected systems). +file( + GLOB_RECURSE THRUST_HEADER_EXCLUDE_DETAILS + RELATIVE ${PROJECT_SOURCE_DIR}/thrust + ${CMAKE_CONFIGURE_DEPENDS} + ${THRUST_HEADER_EXCLUDE_DETAILS_GLOBS} +) +list(REMOVE_ITEM THRUST_HEADERS ${THRUST_HEADER_EXCLUDE_DETAILS}) + +# List of headers that aren't implemented for all backends, but are implemented for CUDA. +set(THRUST_PARTIALLY_IMPLEMENTED_HEADERS_CUDA + async/copy.h + async/for_each.h + async/reduce.h + async/sort.h + async/transform.h + event.h + future.h +) + +# List of headers that aren't implemented for all backends, but are implemented for CPP. +set(THRUST_PARTIALLY_IMPLEMENTED_HEADERS_CPP +) + +# List of headers that aren't implemented for all backends, but are implemented for TBB. +set(THRUST_PARTIALLY_IMPLEMENTED_HEADERS_TBB +) + +# List of headers that aren't implemented for all backends, but are implemented for OMP. +set(THRUST_PARTIALLY_IMPLEMENTED_HEADERS_OMP +) + +# List of all partially implemented headers. +set(THRUST_PARTIALLY_IMPLEMENTED_HEADERS + emptylistguard + ${THRUST_PARTIALLY_IMPLEMENTED_HEADERS_CUDA} + ${THRUST_PARTIALLY_IMPLEMENTED_HEADERS_CPP} + ${THRUST_PARTIALLY_IMPLEMENTED_HEADERS_TBB} + ${THRUST_PARTIALLY_IMPLEMENTED_HEADERS_OMP} +) + +list(REMOVE_DUPLICATES THRUST_PARTIALLY_IMPLEMENTED_HEADERS) + +foreach (THRUST_HEADER IN LISTS THRUST_HEADERS) + if ("${THRUST_HEADER}" IN_LIST THRUST_PARTIALLY_IMPLEMENTED_HEADERS) + # This header is partially implemented on _some_ backends... + if (NOT "${THRUST_HEADER}" IN_LIST THRUST_PARTIALLY_IMPLEMENTED_HEADERS_${THRUST_DEVICE_SYSTEM}) + # ...but not on the selected one. + continue() + endif () + endif () + + set(THRUST_HEADER_TEST_EXT .cpp) + if ("CUDA" STREQUAL "${THRUST_DEVICE_SYSTEM}") + set(THRUST_HEADER_TEST_EXT .cu) + endif () + + set(SOURCE_NAME headers/${THRUST_HEADER}${THRUST_HEADER_TEST_EXT}) + configure_file(cmake/header_test.in ${SOURCE_NAME}) + + list(APPEND THRUST_HEADER_TEST_SOURCES ${SOURCE_NAME}) +endforeach () + +add_library(header-test OBJECT ${THRUST_HEADER_TEST_SOURCES}) +target_include_directories( + header-test + PUBLIC ${PROJECT_SOURCE_DIR} ${CUB_INCLUDE_DIR} +) + +include(CTest) +enable_testing() + +# Handle tests. + +option(THRUST_ENABLE_TESTS_WITH_RDC "Also build all tests with RDC." OFF) + +set(THRUST_TEST_RUN_ARGUMENTS + -DTHRUST_SOURCE=${CMAKE_SOURCE_DIR} + -P "${CMAKE_SOURCE_DIR}/cmake/run_test.cmake") + +list(APPEND THRUST_TESTFRAMEWORK_FILES testing/unittest/testframework.cu) +if ("CUDA" STREQUAL "${THRUST_DEVICE_SYSTEM}") + list(APPEND THRUST_TESTFRAMEWORK_FILES testing/unittest/cuda/testframework.cu) +else () + # When CUDA is disabled, explain to CMake that testframework.cu is actually a C++ file. + set_source_files_properties(testing/unittest/testframework.cu + PROPERTIES + LANGUAGE CXX + COMPILE_FLAGS "${THRUST_TREAT_FILE_AS_CXX}") +endif () + +add_library(thrust_testframework STATIC ${THRUST_TESTFRAMEWORK_FILES}) +target_include_directories( + thrust_testframework + PUBLIC ${PROJECT_SOURCE_DIR} ${CUB_INCLUDE_DIR} + PRIVATE ${PROJECT_SOURCE_DIR}/testing +) + +list(APPEND THRUST_TEST_GLOBS testing/*.cu) +list(APPEND THRUST_TEST_GLOBS testing/*.cpp) + +if ("CUDA" STREQUAL "${THRUST_DEVICE_SYSTEM}") + list(APPEND THRUST_TEST_GLOBS testing/cuda/*.cu) +elseif ("CPP" STREQUAL "${THRUST_DEVICE_SYSTEM}") + list(APPEND THRUST_TEST_GLOBS testing/cpp/*.cu) + list(APPEND THRUST_TEST_GLOBS testing/cpp/*.cpp) +elseif ("OMP" STREQUAL "${THRUST_DEVICE_SYSTEM}") + list(APPEND THRUST_TEST_GLOBS testing/omp/*.cu) + list(APPEND THRUST_TEST_GLOBS testing/omp/*.cpp) +endif () rocm_create_package( NAME rocthrust @@ -207,7 +510,7 @@ foreach (THRUST_TEST_SOURCE IN LISTS THRUST_TESTS) target_include_directories( ${THRUST_TEST} - PUBLIC ${PROJECT_SOURCE_DIR} ${PROJECT_SOURCE_DIR}/dependencies/cub + PUBLIC ${PROJECT_SOURCE_DIR} ${CUB_INCLUDE_DIR} PRIVATE ${PROJECT_SOURCE_DIR}/testing ) @@ -234,7 +537,7 @@ foreach (THRUST_TEST_SOURCE IN LISTS THRUST_TESTS) target_include_directories( ${THRUST_TEST_RDC} - PUBLIC ${PROJECT_SOURCE_DIR} ${PROJECT_SOURCE_DIR}/dependencies/cub + PUBLIC ${PROJECT_SOURCE_DIR} ${CUB_INCLUDE_DIR} PRIVATE ${PROJECT_SOURCE_DIR}/testing ) @@ -333,7 +636,7 @@ foreach (THRUST_EXAMPLE_SOURCE IN LISTS THRUST_EXAMPLES) target_include_directories( ${THRUST_EXAMPLE} - PUBLIC ${PROJECT_SOURCE_DIR} ${PROJECT_SOURCE_DIR}/dependencies/cub + PUBLIC ${PROJECT_SOURCE_DIR} ${CUB_INCLUDE_DIR} PRIVATE ${PROJECT_SOURCE_DIR}/examples ) @@ -356,7 +659,7 @@ foreach (THRUST_EXAMPLE_SOURCE IN LISTS THRUST_EXAMPLES) target_include_directories( ${THRUST_EXAMPLE_RDC} - PUBLIC ${PROJECT_SOURCE_DIR} ${PROJECT_SOURCE_DIR}/dependencies/cub + PUBLIC ${PROJECT_SOURCE_DIR} ${CUB_INCLUDE_DIR} PRIVATE ${PROJECT_SOURCE_DIR}/examples ) diff --git a/dependencies/cub b/dependencies/cub index b2e64cf0f..66a3f9324 160000 --- a/dependencies/cub +++ b/dependencies/cub @@ -1 +1 @@ -Subproject commit b2e64cf0fb4ea7ace6c86ca6765ca7c1087ef82e +Subproject commit 66a3f9324e9cfde6f9f68512c7bb38dff43cd2e1 diff --git a/doc/thrust.dox b/doc/thrust.dox index b74f436f5..95ec1a480 100644 --- a/doc/thrust.dox +++ b/doc/thrust.dox @@ -2063,8 +2063,6 @@ PREDEFINED = THRUST_NOEXCEPT=noexcept \ "THRUST_MR_DEFAULT_ALIGNMENT=alignof(max_align_t)" \ "THRUST_FINAL=final" \ "THRUST_OVERRIDE=" \ - "THRUST_BEGIN_NS=namespace thrust {" \ - "THRUST_END_NS=}" \ "cuda_cub=system::cuda" # If the MACRO_EXPANSION and EXPAND_ONLY_PREDEF tags are set to YES then this diff --git a/examples/cuda/async_reduce.cu b/examples/cuda/async_reduce.cu new file mode 100644 index 000000000..845fe882d --- /dev/null +++ b/examples/cuda/async_reduce.cu @@ -0,0 +1,78 @@ +#include +#include +#include +#include +#include + +#if THRUST_CPP_DIALECT >= 2011 +#include +#endif + +// This example demonstrates two ways to achieve algorithm invocations that are asynchronous with +// the calling thread. +// +// The first method wraps a call to thrust::reduce inside a __global__ function. Since __global__ function +// launches are asynchronous with the launching thread, this achieves asynchrony. The result of the reduction +// is stored to a pointer to CUDA global memory. The calling thread waits for the result of the reduction to +// be ready by synchronizing with the CUDA stream on which the __global__ function is launched. +// +// The second method uses the C++11 library function, std::async, to create concurrency. The lambda function +// given to std::async returns the result of thrust::reduce to a std::future. The calling thread can use the +// std::future to wait for the result of the reduction. This method requires a compiler which supports +// C++11-capable language and library constructs. + +template +__global__ void reduce_kernel(Iterator first, Iterator last, T init, BinaryOperation binary_op, Pointer result) +{ + *result = thrust::reduce(thrust::cuda::par, first, last, init, binary_op); +} + +int main() +{ + size_t n = 1 << 20; + thrust::device_vector data(n, 1); + thrust::device_vector result(1, 0); + + // method 1: call thrust::reduce from an asynchronous CUDA kernel launch + + // create a CUDA stream + cudaStream_t s; + cudaStreamCreate(&s); + + // launch a CUDA kernel with only 1 thread on our stream + reduce_kernel<<<1,1,0,s>>>(data.begin(), data.end(), 0, thrust::plus(), result.data()); + + // wait for the stream to finish + cudaStreamSynchronize(s); + + // our result should be ready + assert(result[0] == n); + + cudaStreamDestroy(s); + + // reset the result + result[0] = 0; + +#if THRUST_CPP_DIALECT >= 2011 + // method 2: use std::async to create asynchrony + + // copy all the algorithm parameters + auto begin = data.begin(); + auto end = data.end(); + unsigned int init = 0; + auto binary_op = thrust::plus(); + + // std::async captures the algorithm parameters by value + // use std::launch::async to ensure the creation of a new thread + std::future future_result = std::async(std::launch::async, [=] + { + return thrust::reduce(begin, end, init, binary_op); + }); + + // wait on the result and check that it is correct + assert(future_result.get() == n); +#endif + + return 0; +} + diff --git a/examples/cuda/global_device_vector.cu b/examples/cuda/global_device_vector.cu index 1419cae62..a99566796 100644 --- a/examples/cuda/global_device_vector.cu +++ b/examples/cuda/global_device_vector.cu @@ -1,3 +1,4 @@ +#include #include // If you create a global `thrust::device_vector` with the default allocator, @@ -20,7 +21,7 @@ typedef thrust::system::cuda::detail::cuda_memory_resource< thrust::cuda::pointer > device_ignore_shutdown_memory_resource; -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 template using device_ignore_shutdown_allocator = thrust::mr::stateless_resource_allocator< diff --git a/internal/benchmark/bench.cu b/internal/benchmark/bench.cu index 5ac4becc3..10f3b2cad 100644 --- a/internal/benchmark/bench.cu +++ b/internal/benchmark/bench.cu @@ -4,7 +4,7 @@ #include #include #include -#include +#include #if THRUST_CPP_DIALECT >= 2011 #include @@ -68,7 +68,7 @@ // We don't use THRUST_NOEXCEPT because it's new, and we want this benchmark to // be backwards-compatible to older versions of Thrust. -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 #define NOEXCEPT noexcept #else #define NOEXCEPT throw() diff --git a/internal/build/common_compiler.mk b/internal/build/common_compiler.mk index 2d6fc28a2..b337c4fe9 100644 --- a/internal/build/common_compiler.mk +++ b/internal/build/common_compiler.mk @@ -117,5 +117,16 @@ else ifeq ($(OS),win32) # Disable warning about applying unary - to unsigned type. CUDACC_FLAGS += -Xcompiler "/wd4146" + + # Warning about declspec(allocator) on inappropriate function types + CUDACC_FLAGS += -Xcompiler "/wd4494" + + # Allow tests to have lots and lots of sections in each translation unit: + CUDACC_FLAGS += -Xcompiler "/bigobj" endif +# Promote all NVCC warnings into errors +CUDACC_FLAGS += -Werror all-warnings + +# Print warning numbers with cudafe diagnostics +CUDACC_FLAGS += -Xcudafe --display_error_number diff --git a/testing/alignment.cu b/testing/alignment.cu index 6ddf1c73c..e55df2e96 100644 --- a/testing/alignment.cu +++ b/testing/alignment.cu @@ -210,7 +210,7 @@ void test_aligned_type() DECLARE_UNITTEST(test_aligned_type); template -void test_aligned_storage_instantiation() +void test_aligned_storage_instantiation(thrust::detail::true_type /* Align is valid */) { typedef typename thrust::detail::aligned_storage::type type; ASSERT_GEQUAL(sizeof(type), Len); @@ -218,6 +218,21 @@ void test_aligned_storage_instantiation() ASSERT_EQUAL(thrust::detail::alignment_of::value, Align); } +template +void test_aligned_storage_instantiation(thrust::detail::false_type /* Align is invalid */) +{ + // no-op -- alignment is > max_align_t and MSVC complains loudly. +} + +template +void test_aligned_storage_instantiation() +{ + typedef thrust::detail::integral_constant< + bool, Align <= THRUST_ALIGNOF(thrust::detail::max_align_t)> + ValidAlign; + test_aligned_storage_instantiation(ValidAlign()); +} + template void test_aligned_storage_size() { diff --git a/testing/allocator.cu b/testing/allocator.cu index 709e04fdb..1b9ee48a5 100644 --- a/testing/allocator.cu +++ b/testing/allocator.cu @@ -16,6 +16,7 @@ */ #include +#include #include #include #include @@ -222,7 +223,7 @@ void TestAllocatorTraitsRebind() } DECLARE_UNITTEST(TestAllocatorTraitsRebind); -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 void TestAllocatorTraitsRebindCpp11() { ASSERT_EQUAL( @@ -270,5 +271,5 @@ void TestAllocatorTraitsRebindCpp11() ); } DECLARE_UNITTEST(TestAllocatorTraitsRebindCpp11); -#endif +#endif // C++11 diff --git a/testing/async_sort.cu b/testing/async_sort.cu index 626e21c3c..c9ae1dd34 100644 --- a/testing/async_sort.cu +++ b/testing/async_sort.cu @@ -1,6 +1,8 @@ #include -#if THRUST_CPP_DIALECT >= 2011 && !defined(THRUST_LEGACY_GCC) +// Disabled on MSVC for GH issue #1098 +#if THRUST_CPP_DIALECT >= 2011 && !defined(THRUST_LEGACY_GCC) && \ + THRUST_HOST_COMPILER != THRUST_HOST_COMPILER_MSVC #include diff --git a/testing/complex.cu b/testing/complex.cu index f8d14941e..175e142ad 100644 --- a/testing/complex.cu +++ b/testing/complex.cu @@ -18,6 +18,8 @@ #include #include +#include + #include #include #include @@ -290,7 +292,7 @@ struct TestComplexTrigonometricFunctions ASSERT_ALMOST_EQUAL(sinh(a),sinh(c)); ASSERT_ALMOST_EQUAL(tanh(a),tanh(c)); -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 ASSERT_ALMOST_EQUAL(acos(a),acos(c)); ASSERT_ALMOST_EQUAL(asin(a),asin(c)); diff --git a/testing/dependencies_aware_policies.cu b/testing/dependencies_aware_policies.cu index b4195c7c9..10aa7f916 100644 --- a/testing/dependencies_aware_policies.cu +++ b/testing/dependencies_aware_policies.cu @@ -1,5 +1,6 @@ #include +#include #include #if THRUST_DEVICE_COMPILER == THRUST_DEVICE_COMPILER_HIP @@ -16,7 +17,7 @@ # include #endif -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 template struct test_allocator_t @@ -187,11 +188,11 @@ SimpleUnitTest< > > TestDependencyAttachmentInstance; -#else +#else // C++11 void TestDummy() { } DECLARE_UNITTEST(TestDummy); -#endif +#endif // C++11 diff --git a/testing/mr_disjoint_pool.cu b/testing/mr_disjoint_pool.cu index 883250671..8499c6c53 100644 --- a/testing/mr_disjoint_pool.cu +++ b/testing/mr_disjoint_pool.cu @@ -1,8 +1,10 @@ #include + +#include #include #include -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 #include #endif @@ -177,7 +179,7 @@ void TestDisjointUnsynchronizedPool() } DECLARE_UNITTEST(TestDisjointUnsynchronizedPool); -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 void TestDisjointSynchronizedPool() { TestDisjointPool(); @@ -260,7 +262,7 @@ void TestDisjointUnsynchronizedPoolCachingOversized() } DECLARE_UNITTEST(TestDisjointUnsynchronizedPoolCachingOversized); -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 void TestDisjointSynchronizedPoolCachingOversized() { TestDisjointPoolCachingOversized(); @@ -285,7 +287,7 @@ void TestUnsynchronizedDisjointGlobalPool() } DECLARE_UNITTEST(TestUnsynchronizedDisjointGlobalPool); -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 void TestSynchronizedDisjointGlobalPool() { TestDisjointGlobalPool(); diff --git a/testing/mr_pool.cu b/testing/mr_pool.cu index bd91c04ea..75b18f038 100644 --- a/testing/mr_pool.cu +++ b/testing/mr_pool.cu @@ -1,8 +1,10 @@ #include + +#include #include #include -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 #include #endif @@ -241,7 +243,7 @@ void TestUnsynchronizedPool() } DECLARE_UNITTEST(TestUnsynchronizedPool); -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 void TestSynchronizedPool() { TestPool(); @@ -324,7 +326,7 @@ void TestUnsynchronizedPoolCachingOversized() } DECLARE_UNITTEST(TestUnsynchronizedPoolCachingOversized); -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 void TestSynchronizedPoolCachingOversized() { TestPoolCachingOversized(); @@ -348,7 +350,7 @@ void TestUnsynchronizedGlobalPool() } DECLARE_UNITTEST(TestUnsynchronizedGlobalPool); -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 void TestSynchronizedGlobalPool() { TestGlobalPool(); diff --git a/testing/vector.cu b/testing/vector.cu index fa8a359f4..1192ab0af 100644 --- a/testing/vector.cu +++ b/testing/vector.cu @@ -16,8 +16,11 @@ */ #include + +#include #include #include + #include #include #include @@ -759,7 +762,7 @@ void TestVectorReversed(void) } DECLARE_VECTOR_UNITTEST(TestVectorReversed); -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 template void TestVectorMove(void) { diff --git a/testing/vector_allocators.cu b/testing/vector_allocators.cu index 00535d1b0..c7276b28c 100644 --- a/testing/vector_allocators.cu +++ b/testing/vector_allocators.cu @@ -1,4 +1,6 @@ #include + +#include #include #include @@ -23,7 +25,7 @@ public: return *this; } -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 stateful_allocator(stateful_allocator && other) : BaseAlloc(std::move(other)), state(other.state) { @@ -129,7 +131,7 @@ void TestVectorAllocatorConstructors() ASSERT_EQUAL(Alloc::last_allocated, 2); Alloc::last_allocated = 0; -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 // FIXME: uncomment this after the vector_base(vector_base&&, const Alloc&) // is fixed and implemented // Vector v5(std::move(v3), alloc2); @@ -188,7 +190,7 @@ void TestVectorAllocatorPropagateOnCopyAssignmentDevice() } DECLARE_UNITTEST(TestVectorAllocatorPropagateOnCopyAssignmentDevice); -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 template void TestVectorAllocatorPropagateOnMoveAssignment() { diff --git a/thrust/addressof.h b/thrust/addressof.h index 5d4dbf349..1134c759b 100644 --- a/thrust/addressof.h +++ b/thrust/addressof.h @@ -11,7 +11,8 @@ # include #endif -THRUST_BEGIN_NS +namespace thrust +{ /////////////////////////////////////////////////////////////////////////////// @@ -28,5 +29,5 @@ T* addressof(T& arg) /////////////////////////////////////////////////////////////////////////////// -THRUST_END_NS +} // end namespace thrust diff --git a/thrust/allocate_unique.h b/thrust/allocate_unique.h index 5daec97e0..8b1562b0e 100644 --- a/thrust/allocate_unique.h +++ b/thrust/allocate_unique.h @@ -18,7 +18,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ // wg21.link/p0316r0 @@ -437,7 +438,7 @@ uninitialized_allocate_unique_n( /////////////////////////////////////////////////////////////////////////////// -THRUST_END_NS +} // end namespace thrust #endif // THRUST_CPP_DIALECT >= 2011 diff --git a/thrust/async/copy.h b/thrust/async/copy.h index 6f603db80..fd2ce239e 100644 --- a/thrust/async/copy.h +++ b/thrust/async/copy.h @@ -33,7 +33,8 @@ #include -THRUST_BEGIN_NS +namespace thrust +{ namespace async { @@ -146,6 +147,6 @@ THRUST_INLINE_CONSTANT copy_detail::copy_fn copy{}; } // namespace async -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/async/for_each.h b/thrust/async/for_each.h index 3bd86a692..fc1814bdc 100644 --- a/thrust/async/for_each.h +++ b/thrust/async/for_each.h @@ -33,7 +33,8 @@ #include -THRUST_BEGIN_NS +namespace thrust +{ namespace async { @@ -113,7 +114,7 @@ THRUST_INLINE_CONSTANT for_each_detail::for_each_fn for_each{}; } // namespace async -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/async/reduce.h b/thrust/async/reduce.h index ab63d6224..a37499584 100644 --- a/thrust/async/reduce.h +++ b/thrust/async/reduce.h @@ -35,7 +35,8 @@ #include -THRUST_BEGIN_NS +namespace thrust +{ namespace async { @@ -95,10 +96,11 @@ struct reduce_fn final , typename ForwardIt, typename Sentinel, typename T > __host__ - static auto call( + static auto call4( thrust::detail::execution_policy_base const& exec , ForwardIt&& first, Sentinel&& last , T&& init + , thrust::true_type ) // ADL dispatch. THRUST_DECLTYPE_RETURNS( @@ -116,9 +118,10 @@ struct reduce_fn final > __host__ static auto - call( + call3( thrust::detail::execution_policy_base const& exec , ForwardIt&& first, Sentinel&& last + , thrust::true_type ) // ADL dispatch. THRUST_DECLTYPE_RETURNS( @@ -136,10 +139,12 @@ struct reduce_fn final template __host__ - static auto call(ForwardIt&& first, Sentinel&& last, T&& init, BinaryOp&& op) - THRUST_DECLTYPE_RETURNS_WITH_SFINAE_CONDITION( - (negation>>::value) - , reduce_fn::call( + static auto call4(ForwardIt&& first, Sentinel&& last, + T&& init, + BinaryOp&& op, + thrust::false_type) + THRUST_DECLTYPE_RETURNS( + reduce_fn::call( thrust::detail::select_system( typename iterator_system>::type{} ) @@ -151,10 +156,11 @@ struct reduce_fn final template __host__ - static auto call(ForwardIt&& first, Sentinel&& last, T&& init) - THRUST_DECLTYPE_RETURNS_WITH_SFINAE_CONDITION( - (negation>>::value) - , reduce_fn::call( + static auto call3(ForwardIt&& first, Sentinel&& last, + T&& init, + thrust::false_type) + THRUST_DECLTYPE_RETURNS( + reduce_fn::call( thrust::detail::select_system( typename iterator_system>::type{} ) @@ -164,6 +170,25 @@ struct reduce_fn final ) ) + // MSVC WAR: MSVC gets angsty and eats all available RAM when we try to detect + // if T1 is an execution_policy by using SFINAE. Switching to a static + // dispatch pattern to prevent this. + template + __host__ + static auto call(T1&& t1, T2&& t2, T3&& t3) + THRUST_DECLTYPE_RETURNS( + reduce_fn::call3(THRUST_FWD(t1), THRUST_FWD(t2), THRUST_FWD(t3), + thrust::is_execution_policy>{}) + ) + + template + __host__ + static auto call(T1&& t1, T2&& t2, T3&& t3, T4&& t4) + THRUST_DECLTYPE_RETURNS( + reduce_fn::call4(THRUST_FWD(t1), THRUST_FWD(t2), THRUST_FWD(t3), THRUST_FWD(t4), + thrust::is_execution_policy>{}) + ) + template __host__ static auto call(ForwardIt&& first, Sentinel&& last) @@ -257,11 +282,12 @@ struct reduce_into_fn final , typename T > __host__ - static auto call( + static auto call5( thrust::detail::execution_policy_base const& exec , ForwardIt&& first, Sentinel&& last , OutputIt&& output , T&& init + , thrust::true_type ) // ADL dispatch. THRUST_DECLTYPE_RETURNS( @@ -280,10 +306,11 @@ struct reduce_into_fn final > __host__ static auto - call( + call4( thrust::detail::execution_policy_base const& exec , ForwardIt&& first, Sentinel&& last , OutputIt&& output + , thrust::true_type ) // ADL dispatch. THRUST_DECLTYPE_RETURNS( @@ -305,15 +332,15 @@ struct reduce_into_fn final , typename T, typename BinaryOp > __host__ - static auto call( + static auto call5( ForwardIt&& first, Sentinel&& last , OutputIt&& output , T&& init , BinaryOp&& op + , thrust::false_type ) - THRUST_DECLTYPE_RETURNS_WITH_SFINAE_CONDITION( - (negation>>::value) - , reduce_into_fn::call( + THRUST_DECLTYPE_RETURNS( + reduce_into_fn::call( thrust::detail::select_system( typename iterator_system>::type{} , typename iterator_system>::type{} @@ -330,14 +357,14 @@ struct reduce_into_fn final , typename T > __host__ - static auto call( + static auto call4( ForwardIt&& first, Sentinel&& last , OutputIt&& output , T&& init + , thrust::false_type ) - THRUST_DECLTYPE_RETURNS_WITH_SFINAE_CONDITION( - (negation>>::value) - , reduce_into_fn::call( + THRUST_DECLTYPE_RETURNS( + reduce_into_fn::call( thrust::detail::select_system( typename iterator_system>::type{} , typename iterator_system>::type{} @@ -374,6 +401,27 @@ struct reduce_into_fn final ) ) + // MSVC WAR: MSVC gets angsty and eats all available RAM when we try to detect + // if T1 is an execution_policy by using SFINAE. Switching to a static + // dispatch pattern to prevent this. + template + __host__ + static auto call(T1&& t1, T2&& t2, T3&& t3, T4&& t4) + THRUST_DECLTYPE_RETURNS( + reduce_into_fn::call4( + THRUST_FWD(t1), THRUST_FWD(t2), THRUST_FWD(t3), THRUST_FWD(t4), + thrust::is_execution_policy>{}) + ) + + template + __host__ + static auto call(T1&& t1, T2&& t2, T3&& t3, T4&& t4, T5&& t5) + THRUST_DECLTYPE_RETURNS( + reduce_into_fn::call5( + THRUST_FWD(t1), THRUST_FWD(t2), THRUST_FWD(t3), THRUST_FWD(t4), + THRUST_FWD(t5), thrust::is_execution_policy>{}) + ) + template THRUST_NODISCARD __host__ auto operator()(Args&&... args) const @@ -388,7 +436,7 @@ THRUST_INLINE_CONSTANT reduce_into_detail::reduce_into_fn reduce_into{}; } // namespace async -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/async/sort.h b/thrust/async/sort.h index 5a3ef067a..0b6a55830 100644 --- a/thrust/async/sort.h +++ b/thrust/async/sort.h @@ -35,7 +35,8 @@ #include -THRUST_BEGIN_NS +namespace thrust +{ namespace async { @@ -199,9 +200,10 @@ struct sort_fn final , typename ForwardIt, typename Sentinel > __host__ - static auto call( + static auto call3( thrust::detail::execution_policy_base const& exec , ForwardIt&& first, Sentinel&& last + , thrust::true_type ) THRUST_DECLTYPE_RETURNS( sort_fn::call( @@ -215,10 +217,11 @@ struct sort_fn final template __host__ - static auto call(ForwardIt&& first, Sentinel&& last, StrictWeakOrdering&& comp) - THRUST_DECLTYPE_RETURNS_WITH_SFINAE_CONDITION( - (negation>>::value) - , sort_fn::call( + static auto call3(ForwardIt&& first, Sentinel&& last, + StrictWeakOrdering&& comp, + thrust::false_type) + THRUST_DECLTYPE_RETURNS( + sort_fn::call( thrust::detail::select_system( typename iterator_system>::type{} ) @@ -227,6 +230,17 @@ struct sort_fn final ) ) + // MSVC WAR: MSVC gets angsty and eats all available RAM when we try to detect + // if T1 is an execution_policy by using SFINAE. Switching to a static + // dispatch pattern to prevent this. + template + __host__ + static auto call(T1&& t1, T2&& t2, T3&& t3) + THRUST_DECLTYPE_RETURNS( + sort_fn::call3(THRUST_FWD(t1), THRUST_FWD(t2), THRUST_FWD(t3), + thrust::is_execution_policy>{}) + ) + template __host__ static auto call(ForwardIt&& first, Sentinel&& last) @@ -256,7 +270,7 @@ THRUST_INLINE_CONSTANT sort_detail::sort_fn sort{}; } // namespace async -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/async/transform.h b/thrust/async/transform.h index a1ecba83a..dac572aad 100644 --- a/thrust/async/transform.h +++ b/thrust/async/transform.h @@ -33,7 +33,8 @@ #include -THRUST_BEGIN_NS +namespace thrust +{ namespace async { @@ -129,7 +130,7 @@ THRUST_INLINE_CONSTANT transform_detail::transform_fn transform{}; } // namespace async -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/complex.h b/thrust/complex.h index 07f90c131..e4ef00aa0 100644 --- a/thrust/complex.h +++ b/thrust/complex.h @@ -70,7 +70,7 @@ namespace detail template struct complex_storage; -#if __cplusplus >= 201103L \ +#if THRUST_CPP_DIALECT >= 2011 \ && (THRUST_HOST_COMPILER == THRUST_HOST_COMPILER_GCC) \ && (THRUST_GCC_VERSION >= 40800) // C++11 implementation, excluding GCC 4.7, which doesn't have `alignas`. diff --git a/thrust/detail/alignment.h b/thrust/detail/alignment.h index c787b0a13..89c8afcd8 100644 --- a/thrust/detail/alignment.h +++ b/thrust/detail/alignment.h @@ -25,7 +25,7 @@ #include // For `std::size_t` and `std::max_align_t`. -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 #include // For `std::alignment_of` and `std::aligned_storage`. #endif @@ -43,7 +43,7 @@ namespace detail /// inside of a `__declspec(align(#))` attribute. As a workaround, you can /// assign the result of \p THRUST_ALIGNOF to a variable and pass the variable /// as the argument to `__declspec(align(#))`. -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 #define THRUST_ALIGNOF(x) alignof(x) #else #define THRUST_ALIGNOF(x) __alignof(x) @@ -54,7 +54,7 @@ namespace detail /// expression. /// /// It is an implementation of C++11's \p std::alignment_of. -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 template using alignment_of = std::alignment_of; #else @@ -97,7 +97,7 @@ namespace detail template struct aligned_type; -#if __cplusplus >= 201103L \ +#if THRUST_CPP_DIALECT >= 2011 \ && (THRUST_HOST_COMPILER == THRUST_HOST_COMPILER_GCC) \ && (THRUST_GCC_VERSION >= 40800) // C++11 implementation, excluding GCC 4.7, which doesn't have `alignas`. @@ -161,7 +161,7 @@ struct aligned_type; /// The behavior is undefined if `Len` is 0 or `Align` is not a power of 2. /// /// It is an implementation of C++11's \p std::aligned_storage. -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 template using aligned_storage = std::aligned_storage; #else @@ -184,7 +184,7 @@ struct aligned_type; /// strict (as large) as that of every scalar type. /// /// It is an implementation of C++11's \p std::max_align_t. -#if __cplusplus >= 201103L \ +#if THRUST_CPP_DIALECT >= 2011 \ && (THRUST_HOST_COMPILER == THRUST_HOST_COMPILER_GCC) \ && (THRUST_GCC_VERSION >= 40900) // GCC 4.7 and 4.8 don't have `std::max_align_t`. diff --git a/thrust/detail/allocator/allocator_traits.h b/thrust/detail/allocator/allocator_traits.h index 36f56b8c8..768f74dab 100644 --- a/thrust/detail/allocator/allocator_traits.h +++ b/thrust/detail/allocator/allocator_traits.h @@ -164,7 +164,7 @@ template::value> typedef typename Alloc::template rebind::other type; }; -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 template class Alloc, typename T, typename... Args, typename U> struct rebind_alloc, U, true> diff --git a/thrust/detail/allocator_aware_execution_policy.h b/thrust/detail/allocator_aware_execution_policy.h index 3a6eb071b..28fd54f9b 100644 --- a/thrust/detail/allocator_aware_execution_policy.h +++ b/thrust/detail/allocator_aware_execution_policy.h @@ -83,7 +83,7 @@ struct allocator_aware_execution_policy return typename execute_with_allocator_type::type(alloc); } -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 // just the rvalue overload // perfect forwarding doesn't help, because a const reference has to be turned // into a value by copying for the purpose of storing it in execute_with_allocator diff --git a/thrust/detail/complex/catrig.h b/thrust/detail/complex/catrig.h index 70adf03ff..0b60286db 100644 --- a/thrust/detail/complex/catrig.h +++ b/thrust/detail/complex/catrig.h @@ -48,6 +48,7 @@ #pragma once +#include #include #include #include @@ -588,7 +589,7 @@ inline double real_part_reciprocal(double x, double y) * Re(catanh(z)) = x/|z|^2 + O(x/z^4) * as z -> infinity, uniformly in x */ -#if __cplusplus >= 201103L || !defined _MSC_VER +#if THRUST_CPP_DIALECT >= 2011 || THRUST_HOST_COMPILER != THRUST_HOST_COMPILER_MSVC __host__ __device__ inline complex catanh(complex z) { @@ -601,7 +602,12 @@ complex catanh(complex z) y = z.imag(); ax = fabs(x); ay = fabs(y); - + + // MSVC needs to pull this in from the std namespace +#if THRUST_HOST_COMPILER == THRUST_HOST_COMPILER_MSVC + using std::atanh; +#endif + /* This helps handle many cases. */ if (y == 0 && ax <= 1) return (complex(atanh(x), y)); @@ -752,7 +758,7 @@ inline complex asin(const complex& z){ return detail::complex::casin(z); } -#if __cplusplus >= 201103L || !defined _MSC_VER +#if THRUST_CPP_DIALECT >= 2011 || THRUST_HOST_COMPILER != THRUST_HOST_COMPILER_MSVC template <> __host__ __device__ inline complex atan(const complex& z){ @@ -773,7 +779,7 @@ inline complex asinh(const complex& z){ return detail::complex::casinh(z); } -#if __cplusplus >= 201103L || !defined _MSC_VER +#if THRUST_CPP_DIALECT >= 2011 || THRUST_HOST_COMPILER != THRUST_HOST_COMPILER_MSVC template <> __host__ __device__ inline complex atanh(const complex& z){ diff --git a/thrust/detail/complex/catrigf.h b/thrust/detail/complex/catrigf.h index db04c466a..aa924717a 100644 --- a/thrust/detail/complex/catrigf.h +++ b/thrust/detail/complex/catrigf.h @@ -50,6 +50,7 @@ #include #include +#include #include #include @@ -386,13 +387,13 @@ inline float real_part_reciprocal(float x, float y) return (x / (x * x + y * y) * scale); } -#if __cplusplus >= 201103L || !defined _MSC_VER +#if THRUST_CPP_DIALECT >= 2011 || THRUST_HOST_COMPILER != THRUST_HOST_COMPILER_MSVC __host__ __device__ inline complex catanhf(complex z) { float x, y, ax, ay, rx, ry; - const volatile float pio2_lo = 6.1232339957367659e-17; /* 0x11a62633145c07.0p-106 */ - const float pio2_hi = 1.5707963267948966e0;/* 0x1921fb54442d18.0p-52 */ + const volatile float pio2_lo = 6.1232339957367659e-17f; /* 0x11a62633145c07.0p-106 */ + const float pio2_hi = 1.5707963267948966e0f;/* 0x1921fb54442d18.0p-52 */ x = z.real(); @@ -421,7 +422,7 @@ complex catanhf(complex z) return (complex(real_part_reciprocal(x, y), copysignf(pio2_hi + pio2_lo, y))); - const float SQRT_3_EPSILON = 5.9801995673e-4; /* 0x9cc471.0p-34 */ + const float SQRT_3_EPSILON = 5.9801995673e-4f; /* 0x9cc471.0p-34 */ if (ax < SQRT_3_EPSILON / 2 && ay < SQRT_3_EPSILON / 2) { raise_inexact(); return (z); @@ -467,7 +468,7 @@ inline complex asin(const complex& z){ return detail::complex::casinf(z); } -#if __cplusplus >= 201103L || !defined _MSC_VER +#if THRUST_CPP_DIALECT >= 2011 || THRUST_HOST_COMPILER != THRUST_HOST_COMPILER_MSVC template <> __host__ __device__ inline complex atan(const complex& z){ @@ -488,7 +489,7 @@ inline complex asinh(const complex& z){ return detail::complex::casinhf(z); } -#if __cplusplus >= 201103L || !defined _MSC_VER +#if THRUST_CPP_DIALECT >= 2011 || THRUST_HOST_COMPILER != THRUST_HOST_COMPILER_MSVC template <> __host__ __device__ inline complex atanh(const complex& z){ diff --git a/thrust/detail/config/compiler.h b/thrust/detail/config/compiler.h index 7a5ac8ca9..92d7f1625 100644 --- a/thrust/detail/config/compiler.h +++ b/thrust/detail/config/compiler.h @@ -55,10 +55,11 @@ #define THRUST_DEVICE_COMPILER_HIP 5 // figure out which host compiler we're using -// TODO we should move the definition of THRUST_DEPRECATED out of this logic -#if defined(_MSC_VER) - #define THRUST_HOST_COMPILER THRUST_HOST_COMPILER_MSVC - #define THRUST_DEPRECATED __declspec(deprecated) +// XXX we should move the definition of THRUST_DEPRECATED out of this logic +#if defined(_MSC_VER) +#define THRUST_HOST_COMPILER THRUST_HOST_COMPILER_MSVC +#define THRUST_MSVC_VERSION _MSC_VER +#define THRUST_MSVC_VERSION_FULL _MSC_FULL_VER #elif defined(__clang__) #define THRUST_HOST_COMPILER THRUST_HOST_COMPILER_CLANG #define THRUST_DEPRECATED __attribute__ ((deprecated)) @@ -213,3 +214,5 @@ x; \ THRUST_DISABLE_CLANG_AND_GCC_INITIALIZER_REORDERING_WARNING_END \ /**/ + + diff --git a/thrust/detail/config/config.h b/thrust/detail/config/config.h index 1ffd8545e..9203cf470 100644 --- a/thrust/detail/config/config.h +++ b/thrust/detail/config/config.h @@ -26,6 +26,7 @@ #include #include #include +#include // host_system.h & device_system.h must be #included as early as possible // because other config headers depend on it #include diff --git a/thrust/detail/config/cpp_dialect.h b/thrust/detail/config/cpp_dialect.h index 06cc3f2f1..1a0e8b676 100755 --- a/thrust/detail/config/cpp_dialect.h +++ b/thrust/detail/config/cpp_dialect.h @@ -1,5 +1,5 @@ /* - * Copyright 2018 NVIDIA Corporation + * Copyright 2020 NVIDIA Corporation * * Licensed under the Apache License, Version 2.0 (the "License"); * you may not use this file except in compliance with the License. @@ -14,19 +14,97 @@ * limitations under the License. */ +/*! \file cpp_dialect.h + * \brief Detect the version of the C++ standard used by the compiler. + */ + #pragma once -#if __cplusplus < 201103L - #define THRUST_CPP03 - #define THRUST_CPP_DIALECT 2003 -#elif __cplusplus < 201402L - #define THRUST_CPP11 - #define THRUST_CPP_DIALECT 2011 -#elif __cplusplus < 201703L - #define THRUST_CPP14 - #define THRUST_CPP_DIALECT 2014 -#else - #define THRUST_CPP17 - #define THRUST_CPP_DIALECT 2017 +#include + +// Deprecation warnings may be silenced by defining the following macros. These +// may be combined. +// - THRUST_IGNORE_DEPRECATED_CPP_DIALECT: +// Ignore all deprecated C++ dialects and outdated compilers. +// - THRUST_IGNORE_DEPRECATED_CPP_11: +// Ignore deprecation warnings when compiling with C++11. C++03 and outdated +// compilers will still issue warnings. +// - THRUST_IGNORE_DEPRECATED_COMPILER +// Ignore deprecation warnings when using deprecated compilers. Compiling +// with C++03 and C++11 will still issue warnings. + +#ifdef THRUST_IGNORE_DEPRECATED_CPP_DIALECT +# define THRUST_IGNORE_DEPRECATED_CPP_11 +# define THRUST_IGNORE_DEPRECATED_COMPILER +#endif + +// Define this to override the built-in detection. +#ifndef THRUST_CPP_DIALECT + +// MSVC does not define __cplusplus correctly. _MSVC_LANG is used instead. +// This macro is only defined in MSVC 2015U3+. +# ifdef _MSVC_LANG // Do not replace with THRUST_HOST_COMPILER test (see above) +// MSVC2015 reports C++14 but lacks extended constexpr support. Treat as C++11. +# if THRUST_MSVC_VERSION < 1910 && _MSVC_LANG > 201103L /* MSVC < 2017 && CPP > 2011 */ +# define THRUST_CPLUSPLUS 201103L /* Fix to 2011 */ +# else +# define THRUST_CPLUSPLUS _MSVC_LANG /* We'll trust this for now. */ +# endif // MSVC 2015 C++14 fix +# else +# define THRUST_CPLUSPLUS __cplusplus +# endif + +// Detect current dialect: +# if THRUST_CPLUSPLUS < 201103L +# define THRUST_CPP_DIALECT 2003 +# elif THRUST_CPLUSPLUS < 201402L +# define THRUST_CPP_DIALECT 2011 +# elif THRUST_CPLUSPLUS < 201703L +# define THRUST_CPP_DIALECT 2014 +# elif THRUST_CPLUSPLUS == 201703L +# define THRUST_CPP_DIALECT 2017 +# elif THRUST_CPLUSPLUS > 201703L // unknown, but is higher than 2017. +# define THRUST_CPP_DIALECT 2020 +# endif + +# undef THRUST_CPLUSPLUS // cleanup + +#endif // !THRUST_CPP_DIALECT + +// Define THRUST_COMPILER_DEPRECATION macro: +#if THRUST_HOST_COMPILER == THRUST_HOST_COMPILER_MSVC +# define THRUST_COMP_DEPR_IMPL(msg) \ + __pragma(message(__FILE__ ":" THRUST_COMP_DEPR_IMPL0(__LINE__) ": warning: " #msg)) +# define THRUST_COMP_DEPR_IMPL0(x) THRUST_COMP_DEPR_IMPL1(x) +# define THRUST_COMP_DEPR_IMPL1(x) #x +#else // clang / gcc: +# define THRUST_COMP_DEPR_IMPL(msg) THRUST_COMP_DEPR_IMPL0(GCC warning #msg) +# define THRUST_COMP_DEPR_IMPL0(expr) _Pragma(#expr) +# define THRUST_COMP_DEPR_IMPL1 /* intentionally blank */ +#endif + +#define THRUST_COMPILER_DEPRECATION(REQ, FIX) \ + THRUST_COMP_DEPR_IMPL(Thrust requires REQ. Please FIX. Define THRUST_IGNORE_DEPRECATED_CPP_DIALECT to suppress this message.) + +// Minimum required compiler checks: +#ifndef THRUST_IGNORE_DEPRECATED_COMPILER +# if THRUST_HOST_COMPILER == THRUST_HOST_COMPILER_GCC && THRUST_GCC_VERSION < 50000 + THRUST_COMPILER_DEPRECATION(GCC 5.0, upgrade your compiler); +# endif +# if THRUST_HOST_COMPILER == THRUST_HOST_COMPILER_CLANG && THRUST_CLANG_VERSION < 60000 + THRUST_COMPILER_DEPRECATION(Clang 6.0, upgrade your compiler); +# endif +# if THRUST_HOST_COMPILER == THRUST_HOST_COMPILER_MSVC && THRUST_MSVC_VERSION < 1910 + THRUST_COMPILER_DEPRECATION(MSVC 2017, upgrade your compiler); +# endif +#endif + +#if !defined(THRUST_IGNORE_DEPRECATED_CPP_DIALECT) && THRUST_CPP_DIALECT < 2014 && \ + (THRUST_CPP_DIALECT != 2011 || !defined(THRUST_IGNORE_DEPRECATED_CPP_11)) + THRUST_COMPILER_DEPRECATION(C++14, pass -std=c++14 to your compiler); #endif +#undef THRUST_COMPILER_DEPRECATION +#undef THRUST_COMP_DEPR_IMPL +#undef THRUST_COMP_DEPR_IMPL0 +#undef THRUST_COMP_DEPR_IMPL1 diff --git a/thrust/detail/util/blocking.h b/thrust/detail/config/deprecated.h similarity index 52% rename from thrust/detail/util/blocking.h rename to thrust/detail/config/deprecated.h index 747d9b97b..cd18f3ac9 100644 --- a/thrust/detail/util/blocking.h +++ b/thrust/detail/config/deprecated.h @@ -1,5 +1,5 @@ /* - * Copyright 2008-2013 NVIDIA Corporation + * Copyright 2018-2020 NVIDIA Corporation * * Licensed under the Apache License, Version 2.0 (the "License"); * you may not use this file except in compliance with the License. @@ -14,24 +14,20 @@ * limitations under the License. */ +/*! \file deprecated.h + * \brief Defines the THRUST_DEPRECATED macro + */ #pragma once -//functions to support blocking - -namespace thrust -{ - -namespace detail -{ - -namespace util -{ - - -} // end namespace util - -} // end namespace detail - -} // end namespace thrust - +#include + +#if THRUST_HOST_COMPILER == THRUST_HOST_COMPILER_MSVC +# define THRUST_DEPRECATED __declspec(deprecated) +#elif THRUST_HOST_COMPILER == THRUST_HOST_COMPILER_CLANG +# define THRUST_DEPRECATED __attribute__((deprecated)) +#elif THRUST_HOST_COMPILER == THRUST_HOST_COMPILER_GCC +# define THRUST_DEPRECATED __attribute__((deprecated)) +#else +# define THRUST_DEPRECATED +#endif diff --git a/thrust/detail/config/exec_check_disable.h b/thrust/detail/config/exec_check_disable.h index ee36b6562..114ca3853 100644 --- a/thrust/detail/config/exec_check_disable.h +++ b/thrust/detail/config/exec_check_disable.h @@ -28,7 +28,11 @@ #if defined(__CUDACC__) && !defined(__NVCOMPILER_CUDA__) && \ !(defined(__CUDA__) && defined(__clang__)) -#define __thrust_exec_check_disable__ #pragma nv_exec_check_disable +#if THRUST_HOST_COMPILER == THRUST_HOST_COMPILER_MSVC +#define __thrust_exec_check_disable__ __pragma("nv_exec_check_disable") +#else // MSVC +#define __thrust_exec_check_disable__ _Pragma("nv_exec_check_disable") +#endif // MSVC #else diff --git a/thrust/detail/contiguous_storage.h b/thrust/detail/contiguous_storage.h index 378cfb815..84485e754 100644 --- a/thrust/detail/contiguous_storage.h +++ b/thrust/detail/contiguous_storage.h @@ -19,6 +19,7 @@ #include #include #include +#include namespace thrust { @@ -167,7 +168,7 @@ template __host__ __device__ void propagate_allocator(const contiguous_storage &other); -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 __host__ __device__ void propagate_allocator(contiguous_storage &other); @@ -220,7 +221,7 @@ template __host__ __device__ void propagate_allocator_dispatch(false_type, const contiguous_storage &other); -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 __host__ __device__ void propagate_allocator_dispatch(true_type, contiguous_storage &other); diff --git a/thrust/detail/contiguous_storage.inl b/thrust/detail/contiguous_storage.inl index f889a6c85..06197616a 100644 --- a/thrust/detail/contiguous_storage.inl +++ b/thrust/detail/contiguous_storage.inl @@ -16,6 +16,7 @@ #pragma once +#include #include #include #include @@ -384,7 +385,7 @@ __host__ __device__ propagate_allocator_dispatch(c, other); } // end contiguous_storage::propagate_allocator() -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 template __host__ __device__ void contiguous_storage @@ -519,7 +520,7 @@ __host__ __device__ { } // end contiguous_storage::propagate_allocator() -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 __thrust_exec_check_disable__ template __host__ __device__ diff --git a/thrust/detail/distance.inl b/thrust/detail/distance.inl index 5732a9c25..f12ef204c 100644 --- a/thrust/detail/distance.inl +++ b/thrust/detail/distance.inl @@ -27,6 +27,7 @@ namespace thrust { +__thrust_exec_check_disable__ template inline __host__ __device__ typename thrust::iterator_traits::difference_type diff --git a/thrust/detail/event_error.h b/thrust/detail/event_error.h index 742439b7e..8b7854a4f 100644 --- a/thrust/detail/event_error.h +++ b/thrust/detail/event_error.h @@ -30,7 +30,8 @@ #include -THRUST_BEGIN_NS +namespace thrust +{ enum class event_errc { @@ -159,7 +160,7 @@ inline bool operator<(event_error const& lhs, event_error const& rhs) noexcept return lhs.code() < rhs.code(); } -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/detail/execute_with_allocator.h b/thrust/detail/execute_with_allocator.h index 0b92d12b3..d18a2a064 100644 --- a/thrust/detail/execute_with_allocator.h +++ b/thrust/detail/execute_with_allocator.h @@ -78,7 +78,7 @@ return_temporary_buffer( alloc_traits::deallocate(system.get_allocator(), to_ptr, 0); } -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 template < typename T, diff --git a/thrust/detail/execute_with_dependencies.h b/thrust/detail/execute_with_dependencies.h index 01fb82364..2fa44a8b9 100644 --- a/thrust/detail/execute_with_dependencies.h +++ b/thrust/detail/execute_with_dependencies.h @@ -189,7 +189,7 @@ struct execute_with_allocator_and_dependencies return std::move(dependencies); } - typename std::remove_reference::type& + typename std::add_lvalue_reference::type __host__ get_allocator() { diff --git a/thrust/detail/memory_algorithms.h b/thrust/detail/memory_algorithms.h index 74e863dcc..de0d53de6 100644 --- a/thrust/detail/memory_algorithms.h +++ b/thrust/detail/memory_algorithms.h @@ -8,6 +8,7 @@ #pragma once +#include #include #include #include @@ -17,7 +18,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ /////////////////////////////////////////////////////////////////////////////// @@ -101,7 +103,7 @@ ForwardIt destroy_n(Allocator const& alloc, ForwardIt first, Size n) return first; } -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 template __host__ __device__ void uninitialized_construct( @@ -204,5 +206,5 @@ void uninitialized_construct_n_with_allocator( /////////////////////////////////////////////////////////////////////////////// -THRUST_END_NS +} // end namespace thrust diff --git a/thrust/detail/pointer.h b/thrust/detail/pointer.h index 26876b4ae..1c86028d6 100644 --- a/thrust/detail/pointer.h +++ b/thrust/detail/pointer.h @@ -216,6 +216,12 @@ template __host__ __device__ explicit operator bool() const; #endif + + __host__ __device__ + static derived_type pointer_to(typename thrust::detail::pointer_traits_detail::pointer_to_param::type r) + { + return thrust::detail::pointer_traits::pointer_to(r); + } }; // end pointer // Output stream operator diff --git a/thrust/detail/select_system.h b/thrust/detail/select_system.h index dd07a28d1..b22ceb0e9 100644 --- a/thrust/detail/select_system.h +++ b/thrust/detail/select_system.h @@ -25,7 +25,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace detail { @@ -78,7 +79,7 @@ THRUST_INLINE_CONSTANT select_system_detail::select_system_fn select_system{}; } // detail -THRUST_END_NS +} // end namespace thrust #endif // THRUST_CPP_DIALECT >= 2011 diff --git a/thrust/detail/static_assert.h b/thrust/detail/static_assert.h index 66d7eb70f..52674dcaf 100644 --- a/thrust/detail/static_assert.h +++ b/thrust/detail/static_assert.h @@ -29,7 +29,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace detail { @@ -86,6 +87,6 @@ template struct static_assert_test {}; } // namespace detail -THRUST_END_NS +} // end namespace thrust diff --git a/thrust/detail/tuple_algorithms.h b/thrust/detail/tuple_algorithms.h index ea50c8c98..530de4b3f 100644 --- a/thrust/detail/tuple_algorithms.h +++ b/thrust/detail/tuple_algorithms.h @@ -26,7 +26,8 @@ #include -THRUST_BEGIN_NS +namespace thrust +{ template auto tuple_subset(Tuple&& t, index_sequence) @@ -104,7 +105,7 @@ THRUST_DECLTYPE_RETURNS( ) ); -THRUST_END_NS +} // end namespace thrust #endif // THRUST_CPP_DIALECT >= 2011 diff --git a/thrust/detail/type_traits.h b/thrust/detail/type_traits.h index ad02ba6f9..9bfe60d31 100644 --- a/thrust/detail/type_traits.h +++ b/thrust/detail/type_traits.h @@ -49,17 +49,14 @@ namespace detail // to the C++14 operator(), but we'd like standard traits to interoperate // with our version when tag dispatching. #if THRUST_CPP_DIALECT >= 2011 - constexpr integral_constant() = default; + integral_constant() = default; - constexpr integral_constant(integral_constant const&) = default; + integral_constant(integral_constant const&) = default; - #if THRUST_CPP_DIALECT >= 2014 - constexpr // In C++11, constexpr makes member functions const. - #endif integral_constant& operator=(integral_constant const&) = default; constexpr __host__ __device__ - integral_constant(std::integral_constant) {} + integral_constant(std::integral_constant) noexcept {} #endif THRUST_CONSTEXPR __host__ __device__ operator value_type() const THRUST_NOEXCEPT { return value; } diff --git a/thrust/detail/type_traits/result_of_adaptable_function.h b/thrust/detail/type_traits/result_of_adaptable_function.h index 5d862affd..8f91ff0b2 100644 --- a/thrust/detail/type_traits/result_of_adaptable_function.h +++ b/thrust/detail/type_traits/result_of_adaptable_function.h @@ -20,7 +20,7 @@ #include #include -#if __cplusplus >= 201103L || defined(__cpp_lib_result_of_sfinae) +#if THRUST_CPP_DIALECT >= 2011 || defined(__cpp_lib_result_of_sfinae) // necessary for std::result_of #include #endif @@ -31,7 +31,7 @@ namespace detail { // In the C++11 mode, by default, result_of_adaptable function inheritfrom std::result_of -#if __cplusplus >= 201103L || defined(__cpp_lib_result_of_sfinae) +#if THRUST_CPP_DIALECT >= 2011 || defined(__cpp_lib_result_of_sfinae) template struct result_of_adaptable_function : std::result_of {}; #else /* cxx11 */ diff --git a/thrust/detail/vector_base.h b/thrust/detail/vector_base.h index 49cd07070..eecedfc14 100644 --- a/thrust/detail/vector_base.h +++ b/thrust/detail/vector_base.h @@ -26,6 +26,7 @@ #include #include #include +#include #include #include @@ -106,7 +107,7 @@ template */ vector_base(const vector_base &v, const Alloc &alloc); - #if __cplusplus >= 201103L + #if THRUST_CPP_DIALECT >= 2011 /*! Move constructor moves from another vector_base. * \param v The vector_base to move. */ @@ -123,7 +124,7 @@ template */ vector_base &operator=(const vector_base &v); - #if __cplusplus >= 201103L + #if THRUST_CPP_DIALECT >= 2011 /*! Move assign operator moves from another vector_base. * \param v The vector_base to move. */ diff --git a/thrust/detail/vector_base.inl b/thrust/detail/vector_base.inl index 9d5511e26..2e2331770 100644 --- a/thrust/detail/vector_base.inl +++ b/thrust/detail/vector_base.inl @@ -19,6 +19,7 @@ * \brief Inline file for vector_base.h. */ +#include #include #include #include @@ -110,7 +111,7 @@ template range_init(v.begin(), v.end()); } // end vector_base::vector_base() -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 template vector_base ::vector_base(vector_base &&v) @@ -139,7 +140,7 @@ template return *this; } // end vector_base::operator=() -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 template vector_base & vector_base diff --git a/thrust/device_make_unique.h b/thrust/device_make_unique.h index cb7e7c3b9..939006f27 100644 --- a/thrust/device_make_unique.h +++ b/thrust/device_make_unique.h @@ -32,7 +32,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ /////////////////////////////////////////////////////////////////////////////// @@ -53,6 +54,6 @@ auto device_make_unique(Args&&... args) /////////////////////////////////////////////////////////////////////////////// -THRUST_END_NS +} // end namespace thrust #endif // THRUST_CPP_DIALECT >= 2011 diff --git a/thrust/future.h b/thrust/future.h index 90dcc705d..12bebf8c6 100644 --- a/thrust/future.h +++ b/thrust/future.h @@ -55,7 +55,8 @@ #include __THRUST_DEVICE_SYSTEM_FUTURE_HEADER #undef __THRUST_DEVICE_SYSTEM_FUTURE_HEADER -THRUST_BEGIN_NS +namespace thrust +{ /////////////////////////////////////////////////////////////////////////////// @@ -172,7 +173,7 @@ using thrust::system::__THRUST_DEVICE_SYSTEM_NAMESPACE::when_all; /////////////////////////////////////////////////////////////////////////////// -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/host_vector.h b/thrust/host_vector.h index 047949089..bd97b69de 100644 --- a/thrust/host_vector.h +++ b/thrust/host_vector.h @@ -135,7 +135,7 @@ template > host_vector(const host_vector &v, const Alloc &alloc) :Parent(v,alloc) {} - #if __cplusplus >= 201103L + #if THRUST_CPP_DIALECT >= 2011 /*! Move constructor moves from another host_vector. * \param v The host_vector to move. */ @@ -159,7 +159,7 @@ template > host_vector &operator=(const host_vector &v) { Parent::operator=(v); return *this; } - #if __cplusplus >= 201103L + #if THRUST_CPP_DIALECT >= 2011 /*! Move assign operator moves from another host_vector. * \param v The host_vector to move. */ diff --git a/thrust/iterator/detail/reverse_iterator.inl b/thrust/iterator/detail/reverse_iterator.inl index 5eb9ac5ff..bb96c497f 100644 --- a/thrust/iterator/detail/reverse_iterator.inl +++ b/thrust/iterator/detail/reverse_iterator.inl @@ -47,14 +47,14 @@ template reverse_iterator ::reverse_iterator(reverse_iterator const &r // XXX msvc screws this up -#ifndef _MSC_VER +#if THRUST_HOST_COMPILER != THRUST_HOST_COMPILER_MSVC , typename thrust::detail::enable_if< thrust::detail::is_convertible< OtherBidirectionalIterator, BidirectionalIterator >::value >::type * -#endif // _MSC_VER +#endif // MSVC ) :super_t(r.base()) { diff --git a/thrust/iterator/reverse_iterator.h b/thrust/iterator/reverse_iterator.h index 2ba97d0ac..365bc34d2 100644 --- a/thrust/iterator/reverse_iterator.h +++ b/thrust/iterator/reverse_iterator.h @@ -180,14 +180,14 @@ template reverse_iterator(reverse_iterator const &r // XXX msvc screws this up // XXX remove these guards when we have static_assert -#ifndef _MSC_VER +#if THRUST_HOST_COMPILER != THRUST_HOST_COMPILER_MSVC , typename thrust::detail::enable_if< thrust::detail::is_convertible< OtherBidirectionalIterator, BidirectionalIterator >::value >::type * = 0 -#endif // _MSC_VER +#endif // MSVC ); /*! \cond diff --git a/thrust/limits.h b/thrust/limits.h index 10434a3cf..f83dde9c3 100644 --- a/thrust/limits.h +++ b/thrust/limits.h @@ -9,10 +9,11 @@ #include -THRUST_BEGIN_NS +namespace thrust +{ template struct numeric_limits : std::numeric_limits {}; -THRUST_END_NS +} // end namespace thrust diff --git a/thrust/mr/allocator.h b/thrust/mr/allocator.h index 7645759ea..4c6c32886 100644 --- a/thrust/mr/allocator.h +++ b/thrust/mr/allocator.h @@ -170,12 +170,12 @@ bool operator!=(const allocator & lhs, const allocator & rhs) THRU return !(lhs == rhs); } -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 template using polymorphic_allocator = allocator >; -#else +#else // C++11 template class polymorphic_allocator : public allocator > @@ -190,7 +190,7 @@ class polymorphic_allocator : public allocator= 201703L +#if THRUST_CPP_DIALECT >= 2017 # if __has_include() # define THRUST_MR_STD_MR_HEADER # define THRUST_MR_STD_MR_NS std::pmr diff --git a/thrust/mr/validator.h b/thrust/mr/validator.h index 7f7e12c76..9376ae870 100644 --- a/thrust/mr/validator.h +++ b/thrust/mr/validator.h @@ -27,7 +27,7 @@ namespace mr template struct validator { -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 static_assert( std::is_base_of, MR>::value, "a type used as a memory resource must derive from memory_resource" diff --git a/thrust/optional.h b/thrust/optional.h index 94d10d902..f2d9bb2a7 100644 --- a/thrust/optional.h +++ b/thrust/optional.h @@ -30,7 +30,7 @@ #include #include -#if (defined(_MSC_VER) && _MSC_VER == 1900) +#if (THRUST_HOST_COMPILER == THRUST_HOST_COMPILER_MSVC && _MSC_VER == 1900) #define THRUST_OPTIONAL_MSVC2015 #endif @@ -68,7 +68,8 @@ !defined(__clang__)) #ifndef THRUST_GCC_LESS_8_TRIVIALLY_COPY_CONSTRUCTIBLE_MUTEX #define THRUST_GCC_LESS_8_TRIVIALLY_COPY_CONSTRUCTIBLE_MUTEX -THRUST_BEGIN_NS +namespace thrust +{ namespace detail { template struct is_trivially_copy_constructible : std::is_trivially_copy_constructible{}; @@ -78,7 +79,7 @@ THRUST_BEGIN_NS : std::is_trivially_copy_constructible{}; #endif } -THRUST_END_NS +} // end namespace thrust #endif #define THRUST_OPTIONAL_IS_TRIVIALLY_COPY_CONSTRUCTIBLE(T) \ @@ -94,12 +95,12 @@ THRUST_END_NS #define THRUST_OPTIONAL_IS_TRIVIALLY_DESTRUCTIBLE(T) std::is_trivially_destructible::value #endif -#if __cplusplus > 201103L +#if THRUST_CPP_DIALECT > 2011 #define THRUST_OPTIONAL_CPP14 #endif // constexpr implies const in C++11, not C++14 -#if (__cplusplus == 201103L || defined(THRUST_OPTIONAL_MSVC2015) || \ +#if (THRUST_CPP_DIALECT == 2011 || defined(THRUST_OPTIONAL_MSVC2015) || \ defined(THRUST_OPTIONAL_GCC49)) /// \exclude #define THRUST_OPTIONAL_CPP11_CONSTEXPR @@ -108,7 +109,8 @@ THRUST_END_NS #define THRUST_OPTIONAL_CPP11_CONSTEXPR constexpr #endif -THRUST_BEGIN_NS +namespace thrust +{ #ifndef THRUST_MONOSTATE_INPLACE_MUTEX #define THRUST_MONOSTATE_INPLACE_MUTEX /// \brief Used to represent an optional with no data; essentially a bool @@ -145,7 +147,7 @@ template struct conjunction : std::conditional, B>::type {}; -#if defined(_LIBCPP_VERSION) && __cplusplus == 201103L +#if defined(_LIBCPP_VERSION) && THRUST_CPP_DIALECT == 2011 #define THRUST_OPTIONAL_LIBCXX_MEM_FN_WORKAROUND #endif @@ -288,7 +290,7 @@ using enable_assign_from_other = detail::enable_if_t< !std::is_assignable &>::value && !std::is_assignable &&>::value>; -#ifdef _MSC_VER +#if THRUST_HOST_COMPILER == THRUST_HOST_COMPILER_MSVC // TODO make a version which works with MSVC template struct is_swappable : std::true_type {}; @@ -1997,7 +1999,7 @@ inline constexpr optional make_optional(std::initializer_list il, return optional(in_place, il, std::forward(args)...); } -#if __cplusplus >= 201703L +#if THRUST_CPP_DIALECT >= 2017 template optional(T)->optional; #endif @@ -2827,7 +2829,7 @@ template class optional { T *m_value; }; -THRUST_END_NS +} // end namespace thrust namespace std { // TODO SFINAE diff --git a/thrust/per_device_resource.h b/thrust/per_device_resource.h index 91d4d9a0d..3c0158aee 100644 --- a/thrust/per_device_resource.h +++ b/thrust/per_device_resource.h @@ -28,7 +28,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ /*! Returns a global instance of \p MR for the current device of the provided system. * @@ -98,6 +99,6 @@ class per_device_allocator : public thrust::mr::allocator }; -THRUST_END_NS +} // end namespace thrust #endif // THRUST_CPP_DIALECT >= 2011 diff --git a/thrust/system/cpp/detail/vector.inl b/thrust/system/cpp/detail/vector.inl index 77f8be3bc..55a1fa4ba 100644 --- a/thrust/system/cpp/detail/vector.inl +++ b/thrust/system/cpp/detail/vector.inl @@ -51,7 +51,7 @@ template : super_t(x) {} -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 template vector ::vector(vector &&x) @@ -89,7 +89,7 @@ template return *this; } -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 template vector & vector diff --git a/thrust/system/cpp/vector.h b/thrust/system/cpp/vector.h index 1748f3d6f..9aeb7206b 100644 --- a/thrust/system/cpp/vector.h +++ b/thrust/system/cpp/vector.h @@ -96,7 +96,7 @@ template > */ vector(const vector &x); - #if __cplusplus >= 201103L + #if THRUST_CPP_DIALECT >= 2011 /*! Move constructor moves from over another \p cpp::vector. * \param x The other \p cpp::vector to move from. */ @@ -130,7 +130,7 @@ template > */ vector &operator=(const vector &x); - #if __cplusplus >= 201103L + #if THRUST_CPP_DIALECT >= 2011 /*! Move assignment operator moves from another \p cpp::vector. * \param x The other \p cpp::vector to move from. * \return *this diff --git a/thrust/system/cuda/detail/adjacent_difference.h b/thrust/system/cuda/detail/adjacent_difference.h index ed8d5a4c9..648ddba3e 100644 --- a/thrust/system/cuda/detail/adjacent_difference.h +++ b/thrust/system/cuda/detail/adjacent_difference.h @@ -43,7 +43,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ template __host__ __device__ OutputIterator @@ -530,7 +531,7 @@ adjacent_difference(execution_policy &policy, } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust // #include diff --git a/thrust/system/cuda/detail/assign_value.h b/thrust/system/cuda/detail/assign_value.h index c21bb7773..f6fd987bf 100644 --- a/thrust/system/cuda/detail/assign_value.h +++ b/thrust/system/cuda/detail/assign_value.h @@ -24,7 +24,8 @@ #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { @@ -97,5 +98,5 @@ inline __host__ __device__ } // end cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/async/copy.h b/thrust/system/cuda/detail/async/copy.h index 8083fccd9..8d8779eb1 100644 --- a/thrust/system/cuda/detail/async/copy.h +++ b/thrust/system/cuda/detail/async/copy.h @@ -54,7 +54,8 @@ #include -THRUST_BEGIN_NS +namespace thrust +{ namespace system { namespace cuda { namespace detail { @@ -228,6 +229,10 @@ auto async_copy_n( template < typename FromPolicy, typename ToPolicy , typename ForwardIt, typename OutputIt + // MSVC2015 WAR: doesn't like decltype(...)::value in superclass definition +, typename IsH2DCopy = decltype(is_host_to_device_copy( + std::declval() + , std::declval())) > struct is_buffered_trivially_relocatable_host_to_device_copy : thrust::integral_constant< @@ -238,12 +243,7 @@ struct is_buffered_trivially_relocatable_host_to_device_copy typename iterator_traits::value_type , typename iterator_traits::value_type >::value - && decltype( - is_host_to_device_copy( - std::declval() - , std::declval() - ) - )::value + && IsH2DCopy::value > {}; @@ -333,6 +333,10 @@ auto async_copy_n( template < typename FromPolicy, typename ToPolicy , typename ForwardIt, typename OutputIt + // MSVC2015 WAR: doesn't like decltype(...)::value in superclass definition +, typename IsD2HCopy = decltype(is_device_to_host_copy( + std::declval() + , std::declval())) > struct is_buffered_trivially_relocatable_device_to_host_copy : thrust::integral_constant< @@ -343,12 +347,7 @@ struct is_buffered_trivially_relocatable_device_to_host_copy typename iterator_traits::value_type , typename iterator_traits::value_type >::value - && decltype( - is_device_to_host_copy( - std::declval() - , std::declval() - ) - )::value + && IsD2HCopy::value > {}; @@ -541,7 +540,7 @@ THRUST_DECLTYPE_RETURNS( } // cuda_cub -THRUST_END_NS +} // end namespace thrust #endif // THRUST_DEVICE_COMPILER == THRUST_DEVICE_COMPILER_NVCC diff --git a/thrust/system/cuda/detail/async/customization.h b/thrust/system/cuda/detail/async/customization.h index 651eb287f..4cabe372f 100644 --- a/thrust/system/cuda/detail/async/customization.h +++ b/thrust/system/cuda/detail/async/customization.h @@ -49,7 +49,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace system { namespace cuda { namespace detail { @@ -120,7 +121,7 @@ THRUST_DECLTYPE_RETURNS( }}} // namespace system::cuda::detail -THRUST_END_NS +} // end namespace thrust #endif // THRUST_DEVICE_COMPILER == THRUST_DEVICE_COMPILER_NVCC diff --git a/thrust/system/cuda/detail/async/for_each.h b/thrust/system/cuda/detail/async/for_each.h index a6faf178f..37d998fe2 100644 --- a/thrust/system/cuda/detail/async/for_each.h +++ b/thrust/system/cuda/detail/async/for_each.h @@ -48,7 +48,8 @@ #include -THRUST_BEGIN_NS +namespace thrust +{ namespace system { namespace cuda { namespace detail { @@ -153,7 +154,7 @@ THRUST_DECLTYPE_RETURNS( } // cuda_cub -THRUST_END_NS +} // end namespace thrust #endif // THRUST_DEVICE_COMPILER == THRUST_DEVICE_COMPILER_NVCC diff --git a/thrust/system/cuda/detail/async/reduce.h b/thrust/system/cuda/detail/async/reduce.h index 78edb60db..8d538250e 100644 --- a/thrust/system/cuda/detail/async/reduce.h +++ b/thrust/system/cuda/detail/async/reduce.h @@ -50,7 +50,8 @@ #include -THRUST_BEGIN_NS +namespace thrust +{ namespace system { namespace cuda { namespace detail { @@ -74,7 +75,7 @@ auto async_reduce_n( using pointer = typename thrust::detail::allocator_traits:: - rebind_traits::pointer; + template rebind_traits::pointer; unique_eager_future_promise_pair fp; @@ -346,7 +347,7 @@ THRUST_DECLTYPE_RETURNS( } // cuda_cub -THRUST_END_NS +} // end namespace thrust #endif // THRUST_DEVICE_COMPILER == THRUST_DEVICE_COMPILER_NVCC diff --git a/thrust/system/cuda/detail/async/sort.h b/thrust/system/cuda/detail/async/sort.h index fe1bb35e5..f258a9c2a 100644 --- a/thrust/system/cuda/detail/async/sort.h +++ b/thrust/system/cuda/detail/async/sort.h @@ -54,7 +54,8 @@ #include -THRUST_BEGIN_NS +namespace thrust +{ namespace system { namespace cuda { namespace detail { @@ -518,7 +519,7 @@ THRUST_DECLTYPE_RETURNS( } // cuda_cub -THRUST_END_NS +} // end namespace thrust #endif // THRUST_DEVICE_COMPILER == THRUST_DEVICE_COMPILER_NVCC diff --git a/thrust/system/cuda/detail/async/transform.h b/thrust/system/cuda/detail/async/transform.h index 55cc1997b..44934f4a6 100644 --- a/thrust/system/cuda/detail/async/transform.h +++ b/thrust/system/cuda/detail/async/transform.h @@ -48,7 +48,8 @@ #include -THRUST_BEGIN_NS +namespace thrust +{ namespace system { namespace cuda { namespace detail { @@ -157,7 +158,7 @@ THRUST_DECLTYPE_RETURNS( } // cuda_cub -THRUST_END_NS +} // end namespace thrust #endif // THRUST_DEVICE_COMPILER == THRUST_DEVICE_COMPILER_NVCC diff --git a/thrust/system/cuda/detail/binary_search.h b/thrust/system/cuda/detail/binary_search.h index bcd156ffb..1859824b8 100644 --- a/thrust/system/cuda/detail/binary_search.h +++ b/thrust/system/cuda/detail/binary_search.h @@ -44,7 +44,8 @@ # define BS_SIMPLE #endif -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { namespace __binary_search { @@ -774,7 +775,7 @@ lower_bound(execution_policy& policy, } } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif #endif diff --git a/thrust/system/cuda/detail/copy.h b/thrust/system/cuda/detail/copy.h index 15dd00b41..ef51e4a5b 100644 --- a/thrust/system/cuda/detail/copy.h +++ b/thrust/system/cuda/detail/copy.h @@ -31,7 +31,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ template __host__ __device__ OutputIt @@ -91,7 +92,7 @@ copy_n(cross_system systems, OutputIterator result); } // namespace cuda_ -THRUST_END_NS +} // end namespace thrust @@ -99,7 +100,8 @@ THRUST_END_NS #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { @@ -190,7 +192,7 @@ copy_n(cross_system systems, } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #include #include diff --git a/thrust/system/cuda/detail/copy_if.h b/thrust/system/cuda/detail/copy_if.h index 7cb8a1e25..04f658172 100644 --- a/thrust/system/cuda/detail/copy_if.h +++ b/thrust/system/cuda/detail/copy_if.h @@ -41,7 +41,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ // XXX declare generic copy_if interface // to avoid circulular dependency from thrust/copy.h template @@ -850,7 +851,7 @@ copy_if(execution_policy &policy, } // func copy_if } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #include #endif diff --git a/thrust/system/cuda/detail/core/agent_launcher.h b/thrust/system/cuda/detail/core/agent_launcher.h index b20bd0c00..7788481c7 100644 --- a/thrust/system/cuda/detail/core/agent_launcher.h +++ b/thrust/system/cuda/detail/core/agent_launcher.h @@ -42,7 +42,8 @@ template class ID_impl; template class Foo { ID_impl t;}; #endif -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { namespace core { @@ -1179,5 +1180,5 @@ namespace core { } // namespace core } -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/core/alignment.h b/thrust/system/cuda/detail/core/alignment.h index bf3873efe..1dc21ebce 100644 --- a/thrust/system/cuda/detail/core/alignment.h +++ b/thrust/system/cuda/detail/core/alignment.h @@ -20,7 +20,8 @@ #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { namespace alignment_of_detail { @@ -245,4 +246,4 @@ struct aligned_storage } // end cuda_ -THRUST_END_NS +} // end namespace thrust diff --git a/thrust/system/cuda/detail/core/triple_chevron_launch.h b/thrust/system/cuda/detail/core/triple_chevron_launch.h index 0db1c7036..deeffac9d 100644 --- a/thrust/system/cuda/detail/core/triple_chevron_launch.h +++ b/thrust/system/cuda/detail/core/triple_chevron_launch.h @@ -32,7 +32,8 @@ #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { namespace launcher { @@ -972,4 +973,4 @@ namespace launcher { } // namespace launcher } // namespace cuda_ -THRUST_END_NS +} // end namespace thrust diff --git a/thrust/system/cuda/detail/core/util.h b/thrust/system/cuda/detail/core/util.h index f5561d8b7..a2c87772e 100644 --- a/thrust/system/cuda/detail/core/util.h +++ b/thrust/system/cuda/detail/core/util.h @@ -36,7 +36,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { namespace core { @@ -766,4 +767,4 @@ using core::sm35; using core::sm30; } // namespace cuda_ -THRUST_END_NS +} // end namespace thrust diff --git a/thrust/system/cuda/detail/count.h b/thrust/system/cuda/detail/count.h index 2ed68d7e7..0d8f0c02d 100644 --- a/thrust/system/cuda/detail/count.h +++ b/thrust/system/cuda/detail/count.h @@ -34,7 +34,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { template &policy, } } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/cross_system.h b/thrust/system/cuda/detail/cross_system.h index 56a20daa2..f89f3dba8 100644 --- a/thrust/system/cuda/detail/cross_system.h +++ b/thrust/system/cuda/detail/cross_system.h @@ -30,7 +30,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { template @@ -114,7 +115,12 @@ namespace cuda_cub { ) ) - template + template (), + std::declval()))> THRUST_CONSTEXPR __host__ __device__ auto is_device_to_host_copy( ExecutionPolicy0 const& exec0 @@ -122,28 +128,32 @@ namespace cuda_cub { ) noexcept -> thrust::detail::integral_constant< - bool - , cudaMemcpyDeviceToHost - == decltype(direction_of_copy(exec0, exec1))::value + bool, cudaMemcpyDeviceToHost == Direction::value > { return {}; } - template + template ()))> THRUST_CONSTEXPR __host__ __device__ auto is_device_to_host_copy(ExecutionPolicy const& exec) noexcept -> thrust::detail::integral_constant< - bool - , cudaMemcpyDeviceToHost - == decltype(direction_of_copy(exec))::value + bool, cudaMemcpyDeviceToHost == Direction::value > { return {}; } - template + template (), + std::declval()))> THRUST_CONSTEXPR __host__ __device__ auto is_host_to_device_copy( ExecutionPolicy0 const& exec0 @@ -151,28 +161,32 @@ namespace cuda_cub { ) noexcept -> thrust::detail::integral_constant< - bool - , cudaMemcpyHostToDevice - == decltype(direction_of_copy(exec0, exec1))::value + bool, cudaMemcpyHostToDevice == Direction::value > { return {}; } - template + template ()))> THRUST_CONSTEXPR __host__ __device__ auto is_host_to_device_copy(ExecutionPolicy const& exec) noexcept -> thrust::detail::integral_constant< - bool - , cudaMemcpyHostToDevice - == decltype(direction_of_copy(exec))::value + bool, cudaMemcpyHostToDevice == Direction::value > { return {}; } - template + template (), + std::declval()))> THRUST_CONSTEXPR __host__ __device__ auto is_device_to_device_copy( ExecutionPolicy0 const& exec0 @@ -180,22 +194,21 @@ namespace cuda_cub { ) noexcept -> thrust::detail::integral_constant< - bool - , cudaMemcpyDeviceToDevice - == decltype(direction_of_copy(exec0, exec1))::value + bool, cudaMemcpyDeviceToDevice == Direction::value > { return {}; } - template + template ()))> THRUST_CONSTEXPR __host__ __device__ auto is_device_to_device_copy(ExecutionPolicy const& exec) noexcept -> thrust::detail::integral_constant< - bool - , cudaMemcpyDeviceToDevice - == decltype(direction_of_copy(exec))::value + bool, cudaMemcpyDeviceToDevice == Direction::value > { return {}; @@ -327,5 +340,5 @@ namespace cuda_cub { } } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust diff --git a/thrust/system/cuda/detail/dispatch.h b/thrust/system/cuda/detail/dispatch.h index f391f9131..45b034217 100644 --- a/thrust/system/cuda/detail/dispatch.h +++ b/thrust/system/cuda/detail/dispatch.h @@ -17,6 +17,7 @@ #pragma once #include +#include /** * Dispatch between 32-bit and 64-bit index based versions of the same algorithm @@ -25,7 +26,7 @@ * interfaces, that always deduce the size type from the arguments. */ #define THRUST_INDEX_TYPE_DISPATCH(status, call, count, arguments) \ - if (count <= std::numeric_limits::max()) { \ + if (count <= thrust::detail::integer_traits::const_max) { \ thrust::detail::int32_t THRUST_PP_CAT2(count, _fixed) = count; \ status = call arguments; \ } \ @@ -44,7 +45,7 @@ * necessary for set algorithms. */ #define THRUST_DOUBLE_INDEX_TYPE_DISPATCH(status, call, count1, count2, arguments) \ - if (count1 + count2 <= std::numeric_limits::max()) { \ + if (count1 + count2 <= thrust::detail::integer_traits::const_max) { \ thrust::detail::int32_t THRUST_PP_CAT2(count1, _fixed) = count1; \ thrust::detail::int32_t THRUST_PP_CAT2(count2, _fixed) = count2; \ status = call arguments; \ @@ -66,7 +67,7 @@ * See reduce_n_impl to see an example of how this is meant to be used. */ #define THRUST_INDEX_TYPE_DISPATCH2(status, call_32, call_64, count, arguments) \ - if (count <= std::numeric_limits::max()) { \ + if (count <= thrust::detail::integer_traits::const_max) { \ thrust::detail::int32_t THRUST_PP_CAT2(count, _fixed) = count; \ status = call_32 arguments; \ } \ diff --git a/thrust/system/cuda/detail/equal.h b/thrust/system/cuda/detail/equal.h index 7a995cffd..dd5e7d686 100644 --- a/thrust/system/cuda/detail/equal.h +++ b/thrust/system/cuda/detail/equal.h @@ -32,7 +32,8 @@ #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { template & policy, } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/execution_policy.h b/thrust/system/cuda/detail/execution_policy.h index 0b3af62e3..ee49a60cb 100644 --- a/thrust/system/cuda/detail/execution_policy.h +++ b/thrust/system/cuda/detail/execution_policy.h @@ -38,7 +38,8 @@ #include #endif -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { @@ -94,5 +95,5 @@ using thrust::cuda_cub::execution_policy; } // namespace cuda -THRUST_END_NS +} // end namespace thrust diff --git a/thrust/system/cuda/detail/extrema.h b/thrust/system/cuda/detail/extrema.h index faef53999..40903cd9a 100644 --- a/thrust/system/cuda/detail/extrema.h +++ b/thrust/system/cuda/detail/extrema.h @@ -37,7 +37,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { namespace __extrema { @@ -563,5 +564,5 @@ minmax_element(execution_policy &policy, } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/fill.h b/thrust/system/cuda/detail/fill.h index b5796f399..078e1b378 100644 --- a/thrust/system/cuda/detail/fill.h +++ b/thrust/system/cuda/detail/fill.h @@ -31,7 +31,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { namespace __fill { @@ -89,5 +90,5 @@ fill(execution_policy& policy, } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/find.h b/thrust/system/cuda/detail/find.h index 0371c1cf8..298be0d1a 100644 --- a/thrust/system/cuda/detail/find.h +++ b/thrust/system/cuda/detail/find.h @@ -34,7 +34,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { // XXX forward declare to circumvent circular depedency @@ -66,12 +67,13 @@ find(execution_policy &policy, T const& value); }; // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { namespace __find_if { @@ -211,5 +213,5 @@ find(execution_policy &policy, } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/for_each.h b/thrust/system/cuda/detail/for_each.h index 7a73242ba..542dcf754 100644 --- a/thrust/system/cuda/detail/for_each.h +++ b/thrust/system/cuda/detail/for_each.h @@ -36,7 +36,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { @@ -104,5 +105,5 @@ namespace cuda_cub { } } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/future.inl b/thrust/system/cuda/detail/future.inl index f64da12a9..8715559d8 100644 --- a/thrust/system/cuda/detail/future.inl +++ b/thrust/system/cuda/detail/future.inl @@ -32,7 +32,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ // Forward declaration. struct new_stream_t; @@ -1362,7 +1363,7 @@ THRUST_DECLTYPE_RETURNS(std::move(dependency)) }} // namespace system::cuda -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/gather.h b/thrust/system/cuda/detail/gather.h index e153a857a..31ca3fd56 100644 --- a/thrust/system/cuda/detail/gather.h +++ b/thrust/system/cuda/detail/gather.h @@ -31,7 +31,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { template & policy, } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/generate.h b/thrust/system/cuda/detail/generate.h index e1058c873..df77901e2 100644 --- a/thrust/system/cuda/detail/generate.h +++ b/thrust/system/cuda/detail/generate.h @@ -34,7 +34,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { // for_each functor @@ -85,5 +86,5 @@ generate(execution_policy &policy, } } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/get_value.h b/thrust/system/cuda/detail/get_value.h index a690dcb1f..9fbb0b548 100644 --- a/thrust/system/cuda/detail/get_value.h +++ b/thrust/system/cuda/detail/get_value.h @@ -23,7 +23,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { @@ -92,6 +93,6 @@ inline __host__ __device__ } // end cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/inner_product.h b/thrust/system/cuda/detail/inner_product.h index 4e1cd5a4c..bd6aec606 100644 --- a/thrust/system/cuda/detail/inner_product.h +++ b/thrust/system/cuda/detail/inner_product.h @@ -33,7 +33,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { @@ -89,5 +90,5 @@ inner_product(execution_policy &policy, } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/internal/copy_cross_system.h b/thrust/system/cuda/detail/internal/copy_cross_system.h index fcdd51f51..ab3b4e5bb 100644 --- a/thrust/system/cuda/detail/internal/copy_cross_system.h +++ b/thrust/system/cuda/detail/internal/copy_cross_system.h @@ -40,7 +40,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { namespace __copy { @@ -238,4 +239,4 @@ namespace __copy { } // namespace __copy } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust diff --git a/thrust/system/cuda/detail/internal/copy_device_to_device.h b/thrust/system/cuda/detail/internal/copy_device_to_device.h index 669211d1e..7a6631d90 100644 --- a/thrust/system/cuda/detail/internal/copy_device_to_device.h +++ b/thrust/system/cuda/detail/internal/copy_device_to_device.h @@ -34,7 +34,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { namespace __copy { @@ -59,5 +60,5 @@ namespace __copy { } // namespace __copy } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/iter_swap.h b/thrust/system/cuda/detail/iter_swap.h index ac224c042..353bb1851 100644 --- a/thrust/system/cuda/detail/iter_swap.h +++ b/thrust/system/cuda/detail/iter_swap.h @@ -24,7 +24,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { @@ -61,5 +62,5 @@ void iter_swap(thrust::cuda::execution_policy &, Pointer1 a, Poin } // end cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/make_unsigned_special.h b/thrust/system/cuda/detail/make_unsigned_special.h index 80fd2a2ea..683647cbe 100644 --- a/thrust/system/cuda/detail/make_unsigned_special.h +++ b/thrust/system/cuda/detail/make_unsigned_special.h @@ -16,7 +16,8 @@ #pragma once -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { namespace detail { @@ -37,5 +38,5 @@ namespace detail { } } -THRUST_END_NS +} // end namespace thrust diff --git a/thrust/system/cuda/detail/malloc_and_free.h b/thrust/system/cuda/detail/malloc_and_free.h index ed6cb87b2..5ca231d0b 100644 --- a/thrust/system/cuda/detail/malloc_and_free.h +++ b/thrust/system/cuda/detail/malloc_and_free.h @@ -29,7 +29,8 @@ #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { #ifdef THRUST_CACHING_DEVICE_MALLOC @@ -99,4 +100,4 @@ void free(execution_policy &, Pointer ptr) } // end free() } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust diff --git a/thrust/system/cuda/detail/merge.h b/thrust/system/cuda/detail/merge.h index 0e080a21e..5a223b606 100644 --- a/thrust/system/cuda/detail/merge.h +++ b/thrust/system/cuda/detail/merge.h @@ -43,7 +43,8 @@ j * Copyright (c) 2016, NVIDIA CORPORATION. All rights reserved. #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { namespace __merge { @@ -1013,5 +1014,5 @@ merge_by_key(execution_policy &policy, } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/mismatch.h b/thrust/system/cuda/detail/mismatch.h index 5854be3ac..98c462e84 100644 --- a/thrust/system/cuda/detail/mismatch.h +++ b/thrust/system/cuda/detail/mismatch.h @@ -33,7 +33,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { template & policy, InputIt1 last1, InputIt2 first2); } // namespace cuda_ -THRUST_END_NS +} // end namespace thrust #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { template & policy, } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/par.h b/thrust/system/cuda/detail/par.h index ace0b3957..1e3be070f 100644 --- a/thrust/system/cuda/detail/par.h +++ b/thrust/system/cuda/detail/par.h @@ -38,7 +38,8 @@ #endif -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { template @@ -145,5 +146,5 @@ namespace cuda { using thrust::cuda_cub::par; } // namespace cuda -THRUST_END_NS +} // end namespace thrust diff --git a/thrust/system/cuda/detail/par_to_seq.h b/thrust/system/cuda/detail/par_to_seq.h index f1610b288..22c4e5838 100644 --- a/thrust/system/cuda/detail/par_to_seq.h +++ b/thrust/system/cuda/detail/par_to_seq.h @@ -29,7 +29,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { template @@ -87,4 +88,4 @@ cvt_to_seq(Policy& policy) #endif } // namespace cuda_ -THRUST_END_NS +} // end namespace thrust diff --git a/thrust/system/cuda/detail/parallel_for.h b/thrust/system/cuda/detail/parallel_for.h index 5e2d027fe..17fa7e7a8 100644 --- a/thrust/system/cuda/detail/parallel_for.h +++ b/thrust/system/cuda/detail/parallel_for.h @@ -36,7 +36,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { @@ -173,5 +174,5 @@ parallel_for(execution_policy &policy, } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/partition.h b/thrust/system/cuda/detail/partition.h index 9be3aa4af..c69d02409 100644 --- a/thrust/system/cuda/detail/partition.h +++ b/thrust/system/cuda/detail/partition.h @@ -43,7 +43,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { namespace __partition { @@ -1141,5 +1142,5 @@ is_partitioned(execution_policy &policy, } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/per_device_resource.h b/thrust/system/cuda/detail/per_device_resource.h index 528ac221d..68f7194af 100644 --- a/thrust/system/cuda/detail/per_device_resource.h +++ b/thrust/system/cuda/detail/per_device_resource.h @@ -43,7 +43,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { @@ -64,7 +65,7 @@ MR * get_per_device_resource(execution_policy&) } -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/reduce.h b/thrust/system/cuda/detail/reduce.h index 72b1d9d7b..9fece9718 100644 --- a/thrust/system/cuda/detail/reduce.h +++ b/thrust/system/cuda/detail/reduce.h @@ -46,7 +46,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ // forward declare generic reduce // to circumvent circular dependency @@ -1067,7 +1068,7 @@ reduce(execution_policy &policy, } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #include #include diff --git a/thrust/system/cuda/detail/reduce_by_key.h b/thrust/system/cuda/detail/reduce_by_key.h index 2169881ff..673a64b82 100644 --- a/thrust/system/cuda/detail/reduce_by_key.h +++ b/thrust/system/cuda/detail/reduce_by_key.h @@ -47,7 +47,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ template &policy, } // namespace cuda_ -THRUST_END_NS +} // end namespace thrust #include #include diff --git a/thrust/system/cuda/detail/remove.h b/thrust/system/cuda/detail/remove.h index 2e252c61d..c590a1adf 100644 --- a/thrust/system/cuda/detail/remove.h +++ b/thrust/system/cuda/detail/remove.h @@ -30,7 +30,8 @@ #if THRUST_DEVICE_COMPILER == THRUST_DEVICE_COMPILER_NVCC #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { // in-place @@ -128,5 +129,5 @@ remove_copy(execution_policy &policy, } } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/replace.h b/thrust/system/cuda/detail/replace.h index 27878337c..d2ccb7b24 100644 --- a/thrust/system/cuda/detail/replace.h +++ b/thrust/system/cuda/detail/replace.h @@ -31,7 +31,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { namespace __replace @@ -206,5 +207,5 @@ replace_copy(execution_policy &policy, } } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/reverse.h b/thrust/system/cuda/detail/reverse.h index 4c2ea42ac..955825217 100644 --- a/thrust/system/cuda/detail/reverse.h +++ b/thrust/system/cuda/detail/reverse.h @@ -30,7 +30,8 @@ #if THRUST_DEVICE_COMPILER == THRUST_DEVICE_COMPILER_NVCC #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { template @@ -47,7 +48,7 @@ reverse(execution_policy &policy, ItemsIt last); } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #include #include @@ -55,7 +56,8 @@ THRUST_END_NS #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { template &policy, } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/scan.h b/thrust/system/cuda/detail/scan.h index 654a1b624..4c3cfefec 100644 --- a/thrust/system/cuda/detail/scan.h +++ b/thrust/system/cuda/detail/scan.h @@ -46,7 +46,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ template &exec, OutputIterator result, T init, AssociativeOperator binary_op); -THRUST_END_NS +} // end namespace thrust -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { namespace __scan { @@ -919,7 +921,7 @@ exclusive_scan(execution_policy &policy, }; } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #include diff --git a/thrust/system/cuda/detail/scan_by_key.h b/thrust/system/cuda/detail/scan_by_key.h index fd1784db8..1744c9e8d 100644 --- a/thrust/system/cuda/detail/scan_by_key.h +++ b/thrust/system/cuda/detail/scan_by_key.h @@ -38,7 +38,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { namespace __scan_by_key { @@ -996,7 +997,7 @@ exclusive_scan_by_key(execution_policy &policy, } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #include diff --git a/thrust/system/cuda/detail/scatter.h b/thrust/system/cuda/detail/scatter.h index e3ba3d87d..3ba0a4b74 100644 --- a/thrust/system/cuda/detail/scatter.h +++ b/thrust/system/cuda/detail/scatter.h @@ -31,7 +31,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { template & policy, } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/set_operations.h b/thrust/system/cuda/detail/set_operations.h index 654553a21..38ba1011d 100644 --- a/thrust/system/cuda/detail/set_operations.h +++ b/thrust/system/cuda/detail/set_operations.h @@ -42,7 +42,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { @@ -1993,5 +1994,5 @@ set_union_by_key(execution_policy &policy, } } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/sort.h b/thrust/system/cuda/detail/sort.h index 850b7739a..b9363b41b 100644 --- a/thrust/system/cuda/detail/sort.h +++ b/thrust/system/cuda/detail/sort.h @@ -46,7 +46,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { namespace __merge_sort { @@ -1743,5 +1744,5 @@ stable_sort_by_key( } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/swap_ranges.h b/thrust/system/cuda/detail/swap_ranges.h index c8d56467b..ba3b47d9b 100644 --- a/thrust/system/cuda/detail/swap_ranges.h +++ b/thrust/system/cuda/detail/swap_ranges.h @@ -35,7 +35,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { @@ -102,5 +103,5 @@ swap_ranges(execution_policy &policy, } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/tabulate.h b/thrust/system/cuda/detail/tabulate.h index 2e5316f4c..70b2720d9 100644 --- a/thrust/system/cuda/detail/tabulate.h +++ b/thrust/system/cuda/detail/tabulate.h @@ -34,7 +34,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { namespace __tabulate { @@ -83,5 +84,5 @@ tabulate(execution_policy& policy, } } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/transform.h b/thrust/system/cuda/detail/transform.h index 85e1cf69b..053fe9095 100644 --- a/thrust/system/cuda/detail/transform.h +++ b/thrust/system/cuda/detail/transform.h @@ -35,7 +35,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { @@ -421,5 +422,5 @@ transform(execution_policy &policy, } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/transform_reduce.h b/thrust/system/cuda/detail/transform_reduce.h index 8cfe2ac71..e9a193f24 100644 --- a/thrust/system/cuda/detail/transform_reduce.h +++ b/thrust/system/cuda/detail/transform_reduce.h @@ -32,7 +32,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { template &policy, } } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/transform_scan.h b/thrust/system/cuda/detail/transform_scan.h index 1ebfea506..500152190 100644 --- a/thrust/system/cuda/detail/transform_scan.h +++ b/thrust/system/cuda/detail/transform_scan.h @@ -32,7 +32,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { @@ -138,5 +139,5 @@ transform_exclusive_scan(execution_policy &policy, } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/uninitialized_copy.h b/thrust/system/cuda/detail/uninitialized_copy.h index 71a72c0e9..8d916e33b 100644 --- a/thrust/system/cuda/detail/uninitialized_copy.h +++ b/thrust/system/cuda/detail/uninitialized_copy.h @@ -34,7 +34,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { @@ -111,5 +112,5 @@ uninitialized_copy(execution_policy& policy, } // namespace cuda_ -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/uninitialized_fill.h b/thrust/system/cuda/detail/uninitialized_fill.h index ad990333f..a8f5fa809 100644 --- a/thrust/system/cuda/detail/uninitialized_fill.h +++ b/thrust/system/cuda/detail/uninitialized_fill.h @@ -34,7 +34,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { @@ -109,5 +110,5 @@ uninitialized_fill(execution_policy& policy, } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #endif diff --git a/thrust/system/cuda/detail/unique.h b/thrust/system/cuda/detail/unique.h index d3ac04364..c2aff4c64 100644 --- a/thrust/system/cuda/detail/unique.h +++ b/thrust/system/cuda/detail/unique.h @@ -42,7 +42,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ template &policy, } } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust // #include diff --git a/thrust/system/cuda/detail/unique_by_key.h b/thrust/system/cuda/detail/unique_by_key.h index 880e5d9a9..e20832131 100644 --- a/thrust/system/cuda/detail/unique_by_key.h +++ b/thrust/system/cuda/detail/unique_by_key.h @@ -44,7 +44,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ template &policy, } // namespace cuda_cub -THRUST_END_NS +} // end namespace thrust #include #include diff --git a/thrust/system/cuda/detail/util.h b/thrust/system/cuda/detail/util.h index 7e2ecbf2c..38136e599 100644 --- a/thrust/system/cuda/detail/util.h +++ b/thrust/system/cuda/detail/util.h @@ -34,7 +34,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { @@ -893,4 +894,4 @@ struct counting_iterator_t } // cuda_ -THRUST_END_NS +} // end namespace thrust diff --git a/thrust/system/cuda/detail/vector.inl b/thrust/system/cuda/detail/vector.inl index 38bb58e4a..dfd4c89b5 100644 --- a/thrust/system/cuda/detail/vector.inl +++ b/thrust/system/cuda/detail/vector.inl @@ -48,7 +48,7 @@ template : super_t(x) {} -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 template vector ::vector(vector &&x) @@ -86,7 +86,7 @@ template return *this; } -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 template vector & vector diff --git a/thrust/system/cuda/future.h b/thrust/system/cuda/future.h index 4709f16a2..fc2986f8b 100644 --- a/thrust/system/cuda/future.h +++ b/thrust/system/cuda/future.h @@ -14,7 +14,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace system { namespace cuda { @@ -66,7 +67,7 @@ unique_eager_future_type( thrust::cuda::execution_policy const& ) noexcept; -THRUST_END_NS +} // end namespace thrust #include diff --git a/thrust/system/cuda/memory.h b/thrust/system/cuda/memory.h index f1510549d..cd27e4da6 100644 --- a/thrust/system/cuda/memory.h +++ b/thrust/system/cuda/memory.h @@ -27,7 +27,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace cuda_cub { /*! Allocates an area of memory available to Thrust's cuda system. @@ -140,7 +141,7 @@ using thrust::cuda_cub::free; using thrust::cuda_cub::allocator; } // end cuda -THRUST_END_NS +} // end namespace thrust #include diff --git a/thrust/system/cuda/memory_resource.h b/thrust/system/cuda/memory_resource.h index 2298981f7..9110e0af4 100644 --- a/thrust/system/cuda/memory_resource.h +++ b/thrust/system/cuda/memory_resource.h @@ -30,7 +30,8 @@ #include -THRUST_BEGIN_NS +namespace thrust +{ namespace system { @@ -106,5 +107,5 @@ typedef detail::pinned_memory_resource universal_host_pinned_memory_resource; } // end cuda } // end system -THRUST_END_NS +} // end namespace thrust diff --git a/thrust/system/cuda/vector.h b/thrust/system/cuda/vector.h index bc2e8d65a..707f9ff7f 100644 --- a/thrust/system/cuda/vector.h +++ b/thrust/system/cuda/vector.h @@ -93,7 +93,7 @@ template > */ vector(const vector &x); - #if __cplusplus >= 201103L + #if THRUST_CPP_DIALECT >= 2011 /*! Move constructor moves from over another \p cuda::vector. * \param x The other \p cuda::vector to move from. */ @@ -125,7 +125,7 @@ template > */ vector &operator=(const vector &x); - #if __cplusplus >= 201103L + #if THRUST_CPP_DIALECT >= 2011 /*! Move assignment operator moves from another \p cuda::vector. * \param x The other \p cuda::vector to move from. * \return *this diff --git a/thrust/system/detail/generic/distance.inl b/thrust/system/detail/generic/distance.inl index 5cc697200..930d0844c 100644 --- a/thrust/system/detail/generic/distance.inl +++ b/thrust/system/detail/generic/distance.inl @@ -60,7 +60,7 @@ inline __host__ __device__ } // end detail - +__thrust_exec_check_disable__ template inline __host__ __device__ typename thrust::iterator_traits::difference_type diff --git a/thrust/system/omp/detail/vector.inl b/thrust/system/omp/detail/vector.inl index 2dac743cb..3e08615f8 100644 --- a/thrust/system/omp/detail/vector.inl +++ b/thrust/system/omp/detail/vector.inl @@ -51,7 +51,7 @@ template : super_t(x) {} -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 template vector ::vector(vector &&x) @@ -89,7 +89,7 @@ template return *this; } -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 template vector & vector diff --git a/thrust/system/omp/vector.h b/thrust/system/omp/vector.h index 1fe7845f3..223ce4935 100644 --- a/thrust/system/omp/vector.h +++ b/thrust/system/omp/vector.h @@ -96,7 +96,7 @@ template > */ vector(const vector &x); - #if __cplusplus >= 201103L + #if THRUST_CPP_DIALECT >= 2011 /*! Move constructor moves another \p omp::vector. * \param x The other \p omp::vector to move from. */ @@ -130,7 +130,7 @@ template > */ vector &operator=(const vector &x); - #if __cplusplus >= 201103L + #if THRUST_CPP_DIALECT >= 2011 /*! Move assignment operator moves another \p omp::vector. * \param x The other \p omp::vector to move. * \return *this diff --git a/thrust/system/tbb/detail/vector.inl b/thrust/system/tbb/detail/vector.inl index fe9d72ab0..5d9cb1c09 100644 --- a/thrust/system/tbb/detail/vector.inl +++ b/thrust/system/tbb/detail/vector.inl @@ -51,7 +51,7 @@ template : super_t(x) {} -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 template vector ::vector(vector &&x) @@ -89,7 +89,7 @@ template return *this; } -#if __cplusplus >= 201103L +#if THRUST_CPP_DIALECT >= 2011 template vector & vector diff --git a/thrust/system/tbb/vector.h b/thrust/system/tbb/vector.h index 1a557ed71..9e12cdc09 100644 --- a/thrust/system/tbb/vector.h +++ b/thrust/system/tbb/vector.h @@ -91,7 +91,7 @@ template > */ vector(const vector &x); - #if __cplusplus >= 201103L + #if THRUST_CPP_DIALECT >= 2011 /*! Move constructor use the move semantic over another \p tbb::vector. * \param x The other \p tbb::vector to move from. */ @@ -125,7 +125,7 @@ template > */ vector &operator=(const vector &x); - #if __cplusplus >= 201103L + #if THRUST_CPP_DIALECT >= 2011 /*! Move assignment operator use move semantic over another \p tbb::vector. * \param x The other \p tbb::vector to move from. * \return *this diff --git a/thrust/type_traits/integer_sequence.h b/thrust/type_traits/integer_sequence.h index 4d04653d1..e28e4f95c 100644 --- a/thrust/type_traits/integer_sequence.h +++ b/thrust/type_traits/integer_sequence.h @@ -23,7 +23,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ #if THRUST_CPP_DIALECT >= 2014 @@ -255,7 +256,7 @@ struct integer_sequence_push_back_impl > } // namespace detail -THRUST_END_NS +} // end namespace thrust #endif // THRUST_CPP_DIALECT >= 2011 diff --git a/thrust/type_traits/is_contiguous_iterator.h b/thrust/type_traits/is_contiguous_iterator.h index 9e704dc31..3e075bd28 100644 --- a/thrust/type_traits/is_contiguous_iterator.h +++ b/thrust/type_traits/is_contiguous_iterator.h @@ -28,7 +28,7 @@ #include -#if defined(_MSC_VER) && _MSC_VER < 1916 // MSVC 2017 version 15.9 +#if THRUST_HOST_COMPILER == THRUST_HOST_COMPILER_MSVC && _MSC_VER < 1916 // MSVC 2017 version 15.9 #include #include #include @@ -38,7 +38,8 @@ #endif #endif -THRUST_BEGIN_NS +namespace thrust +{ namespace detail { @@ -84,10 +85,10 @@ struct proclaim_contiguous_iterator : false_type {}; /// ContiguousIterator /// by specializing `thrust::proclaim_contiguous_iterator`. #define THRUST_PROCLAIM_CONTIGUOUS_ITERATOR(Iterator) \ - THRUST_BEGIN_NS \ + namespace thrust { \ template <> \ struct proclaim_contiguous_iterator : ::thrust::true_type {}; \ - THRUST_END_NS \ + } /* end namespace thrust */ \ /**/ /////////////////////////////////////////////////////////////////////////////// @@ -180,5 +181,5 @@ struct is_contiguous_iterator_impl } // namespace detail -THRUST_END_NS +} // end namespace thrust diff --git a/thrust/type_traits/is_execution_policy.h b/thrust/type_traits/is_execution_policy.h index 5412e6c44..3f2f7ef80 100644 --- a/thrust/type_traits/is_execution_policy.h +++ b/thrust/type_traits/is_execution_policy.h @@ -21,7 +21,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ /// Unary metafunction that is \c true if \c T is an \a ExecutionPolicy and /// \c false otherwise. @@ -44,6 +45,6 @@ template constexpr bool is_execution_policy_v = is_execution_policy::value; #endif -THRUST_END_NS +} // end namespace thrust diff --git a/thrust/type_traits/is_operator_less_or_greater_function_object.h b/thrust/type_traits/is_operator_less_or_greater_function_object.h index 4fb53bda5..6efc00223 100644 --- a/thrust/type_traits/is_operator_less_or_greater_function_object.h +++ b/thrust/type_traits/is_operator_less_or_greater_function_object.h @@ -27,7 +27,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace detail { @@ -131,5 +132,5 @@ struct is_operator_greater_function_object_impl > : true_type } // namespace detail -THRUST_END_NS +} // end namespace thrust diff --git a/thrust/type_traits/is_operator_plus_function_object.h b/thrust/type_traits/is_operator_plus_function_object.h index 80481dfb0..0b2ebb107 100644 --- a/thrust/type_traits/is_operator_plus_function_object.h +++ b/thrust/type_traits/is_operator_plus_function_object.h @@ -26,7 +26,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ namespace detail { @@ -72,5 +73,5 @@ struct is_operator_plus_function_object_impl > : true_type {}; } // namespace detail -THRUST_END_NS +} // end namespace thrust diff --git a/thrust/type_traits/is_trivially_relocatable.h b/thrust/type_traits/is_trivially_relocatable.h index b0cc5c2c7..e47405bc6 100644 --- a/thrust/type_traits/is_trivially_relocatable.h +++ b/thrust/type_traits/is_trivially_relocatable.h @@ -23,7 +23,8 @@ #include #endif -THRUST_BEGIN_NS +namespace thrust +{ namespace detail { @@ -124,10 +125,10 @@ struct proclaim_trivially_relocatable : false_type {}; /// Declares that the type \c T is \a TriviallyRelocatable by specializing /// `thrust::proclaim_trivially_relocatable`. #define THRUST_PROCLAIM_TRIVIALLY_RELOCATABLE(T) \ - THRUST_BEGIN_NS \ + namespace thrust { \ template <> \ struct proclaim_trivially_relocatable : ::thrust::true_type {}; \ - THRUST_END_NS \ + } /* end namespace thrust */ \ /**/ /////////////////////////////////////////////////////////////////////////////// @@ -186,7 +187,7 @@ struct is_trivially_relocatable_impl : is_trivially_relocatable_impl {} } // namespace detail -THRUST_END_NS +} // end namespace thrust #if THRUST_DEVICE_SYSTEM == THRUST_DEVICE_SYSTEM_CUDA diff --git a/thrust/type_traits/logical_metafunctions.h b/thrust/type_traits/logical_metafunctions.h index dbcc18382..5f86ee6a8 100644 --- a/thrust/type_traits/logical_metafunctions.h +++ b/thrust/type_traits/logical_metafunctions.h @@ -19,7 +19,8 @@ #include -THRUST_BEGIN_NS +namespace thrust +{ #if THRUST_CPP_DIALECT >= 2017 @@ -172,7 +173,7 @@ constexpr bool negation_value_v = negation_value::value; template struct negation_value : std::integral_constant {}; -THRUST_END_NS +} // end namespace thrust #endif // THRUST_CPP_DIALECT >= 2011 diff --git a/thrust/type_traits/remove_cvref.h b/thrust/type_traits/remove_cvref.h index ef7304478..4079bfe8e 100644 --- a/thrust/type_traits/remove_cvref.h +++ b/thrust/type_traits/remove_cvref.h @@ -19,7 +19,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ #if THRUST_CPP_DIALECT >= 2020 @@ -43,5 +44,5 @@ using remove_cvref_t = typename remove_cvref::type; #endif // THRUST_CPP_DIALECT >= 2020 -THRUST_END_NS +} // end namespace thrust diff --git a/thrust/type_traits/void_t.h b/thrust/type_traits/void_t.h index 8550cc15b..8ab56a3e8 100644 --- a/thrust/type_traits/void_t.h +++ b/thrust/type_traits/void_t.h @@ -26,7 +26,8 @@ # include #endif -THRUST_BEGIN_NS +namespace thrust +{ #if THRUST_CPP_DIALECT >= 2011 @@ -59,5 +60,5 @@ struct voider #endif -THRUST_END_NS +} // end namespace thrust diff --git a/thrust/version.h b/thrust/version.h index 8ab6b38ed..79dadbfa3 100644 --- a/thrust/version.h +++ b/thrust/version.h @@ -73,9 +73,6 @@ */ #define THRUST_PATCH_NUMBER 0 - -// Declare these namespaces here for the purpose of Doxygenating them - /*! \namespace thrust * \brief \p thrust is the top-level namespace which contains all Thrust * functions and types. @@ -84,12 +81,3 @@ namespace thrust { } - -#ifndef THRUST_BEGIN_NS -#define THRUST_BEGIN_NS namespace thrust { -#endif - -#ifndef THRUST_END_NS -#define THRUST_END_NS } -#endif - diff --git a/thrust/zip_function.h b/thrust/zip_function.h index 26a7f43e7..faea59d4c 100644 --- a/thrust/zip_function.h +++ b/thrust/zip_function.h @@ -17,7 +17,8 @@ #include #include -THRUST_BEGIN_NS +namespace thrust +{ /*! \addtogroup function_objects Function Objects * \{ @@ -205,6 +206,6 @@ auto make_zip_function(Function&& fun) -> zip_function