NVIDIA · brycelelbach · Oct 29, 2020 · Sep 30, 2020 · Sep 30, 2020 · Sep 30, 2020
diff --git a/benchmarks/CMakeLists.txt b/benchmarks/CMakeLists.txt
@@ -0,0 +1,38 @@
+cmake_minimum_required(VERSION 3.18)
+
+project(libcudacxx-benchmarks LANGUAGES CXX CUDA)
+
+set(CMAKE_BUILD_TYPE "RelWithDebInfo")
+
+set(GPU_ARCHS "70;72;75;80" CACHE STRING "List of GPU architectures to compile for.")
+
+message(STATUS "Enabled CUDA architectures:${GPU_ARCHS}")
+
+find_package(Threads REQUIRED)
+find_package(OpenMP)
+
+function(ConfigureHostBench BENCH_NAME BENCH_SRC)
+   add_executable("${BENCH_NAME}" "${BENCH_SRC}")
+   target_link_libraries("${BENCH_NAME}" PRIVATE Threads::Threads)
+
+   # TODO: Link against libcudaxx interface target instead
+   target_include_directories("${BENCH_NAME}" PRIVATE
+                                              "${CMAKE_CURRENT_SOURCE_DIR}/../include")
+   if(OpenMP_CXX_FOUND)
+      target_link_libraries("${BENCH_NAME}" PRIVATE OpenMP::OpenMP_CXX)
+   endif()
+endfunction(ConfigureHostBench)
+
+function(ConfigureDeviceBench BENCH_NAME BENCH_SRC)
+   add_executable("${BENCH_NAME}" "${BENCH_SRC}")
+   set_property(TARGET "${BENCH_NAME}" PROPERTY CUDA_ARCHITECTURES "${GPU_ARCHS}")
+   # TODO: Link against libcudaxx interface target instead
+   target_include_directories("${BENCH_NAME}" PRIVATE
+                                              "${CMAKE_CURRENT_SOURCE_DIR}/../include")
+   target_compile_options("${BENCH_NAME}" PRIVATE --expt-extended-lambda --expt-relaxed-constexpr)
+endfunction(ConfigureDeviceBench)
+
+ConfigureHostBench(concurrency_host concurrency.cpp)                                        
+
+ConfigureDeviceBench(concurrency_device concurrency.cu)
+
diff --git a/samples/benchmark.cpp → benchmarks/concurrency.cpp b/samples/benchmark.cpp → benchmarks/concurrency.cpp
diff --git a/samples/benchmark.cu → benchmarks/concurrency.cu b/samples/benchmark.cu → benchmarks/concurrency.cu
diff --git a/samples/CMakeLists.txt b/samples/CMakeLists.txt
@@ -1,58 +1,44 @@
-cmake_minimum_required(VERSION 3.8)
-project(libcudacxx-samples CXX)
-
-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)
+cmake_minimum_required(VERSION 3.18)
+
+project(libcudacxx-examples LANGUAGES CXX CUDA)
 
 set(CMAKE_BUILD_TYPE "RelWithDebInfo")
-set(CMAKE_CUDA_FLAGS "${CMAKE_CUDA_FLAGS} --expt-relaxed-constexpr --expt-extended-lambda")
-
-set(HIGHEST_COMPUTE_ARCH 80)
-set(KNOWN_COMPUTE_ARCHS 70 72 75 80)
-
-option(DISABLE_ARCH_BY_DEFAULT "If ON, then all CUDA architectures are disabled on the initial CMake run." OFF)
-set(OPTION_INIT ON)
-if (DISABLE_ARCH_BY_DEFAULT)
-  set(OPTION_INIT OFF)
-endif ()
-if (NOT ${HIGHEST_COMPUTE_ARCH} IN_LIST 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 KNOWN_COMPUTE_ARCHS)
-  option(ENABLE_COMPUTE_${COMPUTE_ARCH} "Enable code generation for tests for sm_${COMPUTE_ARCH}" ${OPTION_INIT})
-  if (ENABLE_COMPUTE_${COMPUTE_ARCH})
-    set(COMPUTE_ARCHS "${COMPUTE_ARCHS} ${COMPUTE_ARCH}")
-    set(COMPUTE_MESSAGE "${COMPUTE_MESSAGE} sm_${COMPUTE_ARCH}")
-    set(CMAKE_CUDA_FLAGS "${CMAKE_CUDA_FLAGS} -gencode arch=compute_${COMPUTE_ARCH},code=sm_${COMPUTE_ARCH}")
-  endif ()
-endforeach ()
-
-option(ENABLE_COMPUTE_FUTURE "Enable code generation for tests for compute_${HIGHEST_COMPUTE_ARCH}" ${OPTION_INIT})
-if (ENABLE_COMPUTE_FUTURE)
-  set(COMPUTE_MESSAGE "${COMPUTE_MESSAGE} compute_${HIGHEST_COMPUTE_ARCH}")
-  set(CMAKE_CUDA_FLAGS "${CMAKE_CUDA_FLAGS} -gencode arch=compute_${HIGHEST_COMPUTE_ARCH},code=compute_${HIGHEST_COMPUTE_ARCH}")
-endif ()
-
-message(STATUS "Enabled CUDA architectures:${COMPUTE_MESSAGE}")
-
-include_directories(${CMAKE_SOURCE_DIR}/../include)
 
+find_package(CUDAToolkit REQUIRED)
 find_package(Threads REQUIRED)
 find_package(OpenMP)
 
-add_executable(benchmark_host benchmark.cpp)
-
-target_link_libraries(benchmark_host PUBLIC Threads::Threads)
-if(OpenMP_CXX_FOUND)
-    target_link_libraries(benchmark_host PUBLIC OpenMP::OpenMP_CXX)
+add_executable(trie_st trie_st.cpp)
+target_compile_features(trie_st PRIVATE cxx_std_11)
+
+# Download input files for the trie examples.
+if(NOT (EXISTS books))
+    execute_process(COMMAND mkdir books)
+    file(DOWNLOAD https://www.gutenberg.org/files/2600/2600-0.txt books/2600-0.txt SHOW_PROGRESS)
+    file(DOWNLOAD http://www.gutenberg.org/cache/epub/996/pg996.txt books/pg996.txt SHOW_PROGRESS)
+    file(DOWNLOAD http://www.gutenberg.org/cache/epub/55/pg55.txt books/pg55.txt SHOW_PROGRESS)
+    file(DOWNLOAD https://www.gutenberg.org/files/8800/8800.txt books/8800.txt SHOW_PROGRESS)
+    file(DOWNLOAD https://www.gutenberg.org/files/84/84-0.txt books/84-0.txt SHOW_PROGRESS)
+    file(DOWNLOAD http://www.gutenberg.org/cache/epub/6130/pg6130.txt books/pg6130.txt SHOW_PROGRESS)
+    file(DOWNLOAD http://www.gutenberg.org/cache/epub/1727/pg1727.txt books/pg1727.txt SHOW_PROGRESS)
+    file(DOWNLOAD https://www.gutenberg.org/files/2701/2701-0.txt books/2701-0.txt SHOW_PROGRESS)
+    file(DOWNLOAD https://www.gutenberg.org/files/35/35-0.txt books/35-0.txt SHOW_PROGRESS)
+    file(DOWNLOAD https://www.gutenberg.org/files/1342/1342-0.txt books/1342-0.txt SHOW_PROGRESS)
 endif()
 
-add_executable(benchmark_device benchmark.cu)
+add_executable(trie_mt trie_mt.cpp)
+target_compile_features(trie_mt PRIVATE cxx_std_11)
+target_link_libraries(trie_mt Threads::Threads)
+
+if(CUDAToolkit_VERSION VERSION_GREATER_EQUAL 11.1)
+    add_executable(trie_cuda trie.cu)
+    target_compile_features(trie_cuda PRIVATE cxx_std_11 cuda_std_11)
+    target_compile_options(trie_cuda PRIVATE --expt-relaxed-constexpr)
+    set_property(TARGET trie_cuda PROPERTY CUDA_ARCHITECTURES 70)
+else()
+    message(STATUS "Insufficient CUDA version. Skipping trie.cu example.")
+endif()    
+
+add_executable(rtc rtc_example.cpp)
+target_link_libraries(rtc CUDA::nvrtc)
+target_compile_features(rtc PRIVATE cxx_std_11)
diff --git a/samples/books.sh b/samples/books.sh
diff --git a/samples/linux.sh b/samples/linux.sh
diff --git a/samples/linux_clang.sh b/samples/linux_clang.sh
diff --git a/samples/linux_rtc.sh b/samples/linux_rtc.sh
diff --git a/samples/openmp.sh b/samples/openmp.sh
diff --git a/samples/rtc_example.cpp b/samples/rtc_example.cpp
@@ -39,9 +39,9 @@ THE SOFTWARE.
 const char *trie =
 R"xxx(
 
-#include <simt/cstddef>
-#include <simt/cstdint>
-#include <simt/atomic>
+#include <cuda/std/cstddef>
+#include <cuda/std/cstdint>
+#include <cuda/std/atomic>
 
 template<class T> static constexpr T min(T a, T b) { return a < b ? a : b; }
 
@@ -128,14 +128,12 @@ int main(int argc, char *argv[])
 
   const char *opts[] = {"-std=c++11",
                         "-I/usr/include/linux",
-                        "-I/usr/include/c++/7.3.0",
                         "-I/usr/local/cuda/include",
-                        "-I/home/olivier/freestanding/include",
                         "--gpu-architecture=compute_70",
                         "--relocatable-device-code=true",
                         "-default-device"};
   nvrtcResult compileResult = nvrtcCompileProgram(prog,  // prog
-                                                  8,     // numOptions
+                                                  6,     // numOptions
                                                   opts); // options
   // Obtain compilation log from the program.
   size_t logSize;

diff --git a/samples/trie.cu b/samples/trie.cu
@@ -179,13 +179,18 @@ int main() {
     std::basic_string<char, std::char_traits<char>, managed_allocator<char>>  input;
 
     char const* files[] = {
-        "2600-0.txt", "2701-0.txt", "35-0.txt", "84-0.txt", "8800.txt",
-        "pg1727.txt", "pg55.txt", "pg6130.txt", "pg996.txt", "1342-0.txt"
+        "books/2600-0.txt", "books/2701-0.txt", "books/35-0.txt", "books/84-0.txt", "books/8800.txt",
+      	"books/pg1727.txt", "books/pg55.txt", "books/pg6130.txt", "books/pg996.txt", "books/1342-0.txt"
     };
 
     for(auto* ptr : files) {
+        std::cout << ptr << std::endl;
         auto const cur = input.size();
         std::ifstream in(ptr);
+        if(in.fail()) {
+            std::cerr << "Failed to open file: " << ptr << std::endl;
+            return -1;
+        }
         in.seekg(0, std::ios_base::end);
         auto const pos = in.tellg();
         input.resize(cur + pos);

diff --git a/samples/trie_mt.cpp b/samples/trie_mt.cpp
@@ -120,13 +120,18 @@ int main() {
     std::string input;
 
     char const* files[] = {
-        "2600-0.txt", "2701-0.txt", "35-0.txt", "84-0.txt", "8800.txt",
-      	"pg1727.txt", "pg55.txt", "pg6130.txt", "pg996.txt", "1342-0.txt"
+        "books/2600-0.txt", "books/2701-0.txt", "books/35-0.txt", "books/84-0.txt", "books/8800.txt",
+      	"books/pg1727.txt", "books/pg55.txt", "books/pg6130.txt", "books/pg996.txt", "books/1342-0.txt"
     };
 
     for(auto* ptr : files) {
+        std::cout << ptr << std::endl;
         auto const cur = input.size();
         std::ifstream in(ptr);
+        if(in.fail()) {
+            std::cerr << "Failed to open file: " << ptr << std::endl;
+            return -1;
+        }
         in.seekg(0, std::ios_base::end);
         auto const pos = in.tellg();
         input.resize(cur + pos);

diff --git a/samples/trie_st.cpp b/samples/trie_st.cpp
@@ -91,13 +91,18 @@ int main() {
     std::string input;
 
     char const* files[] = {
-        "2600-0.txt", "2701-0.txt", "35-0.txt", "84-0.txt", "8800.txt",
-      	"pg1727.txt", "pg55.txt", "pg6130.txt", "pg996.txt", "1342-0.txt"
+        "books/2600-0.txt", "books/2701-0.txt", "books/35-0.txt", "books/84-0.txt", "books/8800.txt",
+      	"books/pg1727.txt", "books/pg55.txt", "books/pg6130.txt", "books/pg996.txt", "books/1342-0.txt"
     };
 
     for(auto* ptr : files) {
+        std::cout << ptr << std::endl;
         auto const cur = input.size();
         std::ifstream in(ptr);
+        if(in.fail()) {
+            std::cerr << "Failed to open file: " << ptr << std::endl;
+            return -1;
+        }
         in.seekg(0, std::ios_base::end);
         auto const pos = in.tellg();
         input.resize(cur + pos);

diff --git a/samples/windows.bat b/samples/windows.bat