Merge pull request #75 from gunrock/dev

Feature merge (dev to master)

.gitignore

@@ -36,3 +36,6 @@ eval/
 
 # Ignore build directory
 build
+
+# Ignore external fetched content
+externals/*

CMakeLists.txt

@@ -20,7 +20,7 @@ set(PROJECT_DEPS_DIR externals)
 # end /* Dependencies directory */
 
 # begin /* Include cmake modules */
-include(${PROJECT_SOURCE_DIR}/cmake/FetchRapidJSON.cmake)
+# include(${PROJECT_SOURCE_DIR}/cmake/FetchRapidJSON.cmake)
 include(${PROJECT_SOURCE_DIR}/cmake/FetchThrustCUB.cmake)
 include(${PROJECT_SOURCE_DIR}/cmake/FetchModernGPU.cmake)
 # end /* Include cmake modules */
@@ -157,4 +157,4 @@ option(ESSENTIALS_BUILD_TESTS
 # Subdirectories for examples, testing and documentation
 if(ESSENTIALS_BUILD_TESTS)
   add_subdirectory(unittests)
-endif(ESSENTIALS_BUILD_TESTS)
+endif(ESSENTIALS_BUILD_TESTS)

README.md

@@ -14,6 +14,8 @@ make sssp # or for all applications, use: make -j$(nproc)
 bin/sssp ../datasets/chesapeake.mtx
 ```
 
+##### Preferred **CUDA v11.2.1** due to support for stream ordered memory allocators (e.g. `cudaFreeAsync()`).
+
 ## Getting Started with Gunrock
 
 - **Tutorial:** [Gunrock's programming model]()

cmake/FetchModernGPU.cmake

@@ -1,10 +1,16 @@
 include(FetchContent)
+set(FETCHCONTENT_QUIET ON)
+
+message("-- Cloning External Project: ModernGPU")
+get_filename_component(FC_BASE "../externals"
+                REALPATH BASE_DIR "${CMAKE_BINARY_DIR}")
+set(FETCHCONTENT_BASE_DIR ${FC_BASE})
+
 FetchContent_Declare(
     moderngpu
     GIT_REPOSITORY https://github.com/moderngpu/moderngpu.git
     # tag at master branch:
     GIT_TAG        2b3985541c8e88a133769598c406c33ddde9d0a5
-
 )
 
 FetchContent_GetProperties(moderngpu)
@@ -13,4 +19,4 @@ if(NOT moderngpu_POPULATED)
     moderngpu
   )
 endif()
-set(MODERNGPU_INCLUDE_DIR "${moderngpu_SOURCE_DIR}/src")
+set(MODERNGPU_INCLUDE_DIR "${moderngpu_SOURCE_DIR}/src")

cmake/FetchRapidJSON.cmake

@@ -1,4 +1,11 @@
 include(FetchContent)
+set(FETCHCONTENT_QUIET ON)
+
+message("-- Cloning External Project: RapidJSON")
+get_filename_component(FC_BASE "../externals"
+                REALPATH BASE_DIR "${CMAKE_BINARY_DIR}")
+set(FETCHCONTENT_BASE_DIR ${FC_BASE})
+
 FetchContent_Declare(
     rapidjson
     GIT_REPOSITORY https://github.com/Tencent/rapidjson.git
@@ -11,4 +18,4 @@ if(NOT rapidjson_POPULATED)
     rapidjson
   )
 endif()
-set(RAPIDJSON_INCLUDE_DIR "${rapidjson_SOURCE_DIR}/include")
+set(RAPIDJSON_INCLUDE_DIR "${rapidjson_SOURCE_DIR}/include")

cmake/FetchThrustCUB.cmake

@@ -1,4 +1,11 @@
 include(FetchContent)
+set(FETCHCONTENT_QUIET ON)
+
+message("-- Cloning External Project: Thrust")
+get_filename_component(FC_BASE "../externals"
+                REALPATH BASE_DIR "${CMAKE_BINARY_DIR}")
+set(FETCHCONTENT_BASE_DIR ${FC_BASE})
+
 FetchContent_Declare(
     thrust
     GIT_REPOSITORY https://github.com/thrust/thrust.git
@@ -12,4 +19,4 @@ if(NOT thrust_POPULATED)
   )
 endif()
 set(THRUST_INCLUDE_DIR "${thrust_SOURCE_DIR}")
-set(CUB_INCLUDE_DIR "${thrust_SOURCE_DIR}/cub")
+set(CUB_INCLUDE_DIR "${thrust_SOURCE_DIR}/cub")

examples/CMakeLists.txt

@@ -4,4 +4,6 @@ add_subdirectory(bfs)
 add_subdirectory(color)
 add_subdirectory(geo)
 add_subdirectory(pr)
+add_subdirectory(ppr)
+add_subdirectory(bc)
 # end /* Add examples' subdirectories */

examples/bc/CMakeLists.txt

@@ -0,0 +1,21 @@
+# begin /* Set the application name. */
+set(APPLICATION_NAME bc)
+# end /* Set the application name. */
+
+# begin /* Add CUDA executables */
+add_executable(${APPLICATION_NAME})
+
+set(SOURCE_LIST 
+    ${APPLICATION_NAME}.cu
+)
+
+target_sources(${APPLICATION_NAME} PRIVATE ${SOURCE_LIST})
+target_link_libraries(${APPLICATION_NAME} PRIVATE essentials)
+get_target_property(ESSENTIALS_ARCHITECTURES essentials CUDA_ARCHITECTURES)
+set_target_properties(${APPLICATION_NAME} 
+    PROPERTIES 
+        CUDA_ARCHITECTURES ${ESSENTIALS_ARCHITECTURES}
+) # XXX: Find a better way to inherit essentials properties.
+
+message("-- Example Added: ${APPLICATION_NAME}")
+# end /* Add CUDA executables */

examples/bc/bc.cu

@@ -0,0 +1,71 @@
+#include <gunrock/applications/bc.hxx>
+
+using namespace gunrock;
+using namespace memory;
+
+void test_bc(int num_arguments, char** argument_array) {
+  if (num_arguments != 2) {
+    std::cerr << "usage: ./bin/<program-name> filename.mtx" << std::endl;
+    exit(1);
+  }
+
+  // --
+  // Define types
+
+  using vertex_t = int;
+  using edge_t = int;
+  using weight_t = float;
+
+  // --
+  // IO
+
+  std::string filename = argument_array[1];
+
+  io::matrix_market_t<vertex_t, edge_t, weight_t> mm;
+
+  using csr_t =
+      format::csr_t<memory_space_t::device, vertex_t, edge_t, weight_t>;
+  csr_t csr;
+  csr.from_coo(mm.load(filename));
+
+  // --
+  // Build graph
+
+  auto G = graph::build::from_csr<memory_space_t::device, graph::view_t::csr>(
+      csr.number_of_rows,               // rows
+      csr.number_of_columns,            // columns
+      csr.number_of_nonzeros,           // nonzeros
+      csr.row_offsets.data().get(),     // row_offsets
+      csr.column_indices.data().get(),  // column_indices
+      csr.nonzero_values.data().get()   // values
+  );  // supports row_indices and column_offsets (default = nullptr)
+
+  // --
+  // Params and memory allocation
+
+  // vertex_t single_source = 0;
+  vertex_t n_vertices = G.get_number_of_vertices();
+  thrust::device_vector<weight_t> bc_values(n_vertices);
+
+  // --
+  // GPU Run
+
+  float gpu_elapsed =
+      gunrock::bc::run(G, /* single_source, */ bc_values.data().get());
+
+  // --
+  // Log + Validate
+
+  std::cout << "GPU bc_values (output) = ";
+  thrust::copy(bc_values.begin(),
+               (bc_values.size() < 40) ? bc_values.begin() + bc_values.size()
+                                       : bc_values.begin() + 40,
+               std::ostream_iterator<weight_t>(std::cout, " "));
+  std::cout << std::endl;
+
+  std::cout << "GPU Elapsed Time : " << gpu_elapsed << " (ms)" << std::endl;
+}
+
+int main(int argc, char** argv) {
+  test_bc(argc, argv);
+}

examples/bfs/bfs.cu

@@ -28,24 +28,26 @@ void test_bfs(int num_arguments, char** argument_array) {
       format::csr_t<memory_space_t::device, vertex_t, edge_t, weight_t>;
   csr_t csr;
   csr.from_coo(mm.load(filename));
-  
+
   thrust::device_vector<vertex_t> row_indices(csr.number_of_nonzeros);
+  thrust::device_vector<vertex_t> column_indices(csr.number_of_nonzeros);
   thrust::device_vector<edge_t> column_offsets(csr.number_of_columns + 1);
 
   // --
   // Build graph + metadata
 
-  auto G = graph::build::from_csr<memory_space_t::device,
-                                  graph::view_t::csr | graph::view_t::csc>(
-      csr.number_of_rows,               // rows
-      csr.number_of_columns,            // columns
-      csr.number_of_nonzeros,           // nonzeros
-      csr.row_offsets.data().get(),     // row_offsets
-      csr.column_indices.data().get(),  // column_indices
-      csr.nonzero_values.data().get(),  // values
-      row_indices.data().get(),         // row_indices
-      column_offsets.data().get()       // column_offsets
-  );
+  auto G =
+      graph::build::from_csr<memory_space_t::device,
+                             graph::view_t::csr /* | graph::view_t::csc */>(
+          csr.number_of_rows,               // rows
+          csr.number_of_columns,            // columns
+          csr.number_of_nonzeros,           // nonzeros
+          csr.row_offsets.data().get(),     // row_offsets
+          csr.column_indices.data().get(),  // column_indices
+          csr.nonzero_values.data().get(),  // values
+          row_indices.data().get(),         // row_indices
+          column_offsets.data().get()       // column_offsets
+      );
 
   // --
   // Params and memory allocation
@@ -59,8 +61,8 @@ void test_bfs(int num_arguments, char** argument_array) {
   // --
   // Run problem
 
-  float gpu_elapsed = gunrock::bfs::run(G, single_source, distances.data().get(),
-                                    predecessors.data().get());
+  float gpu_elapsed = gunrock::bfs::run(
+      G, single_source, distances.data().get(), predecessors.data().get());
 
   // --
   // CPU Run
@@ -94,7 +96,6 @@ void test_bfs(int num_arguments, char** argument_array) {
   std::cout << "GPU Elapsed Time : " << gpu_elapsed << " (ms)" << std::endl;
   std::cout << "CPU Elapsed Time : " << cpu_elapsed << " (ms)" << std::endl;
   std::cout << "Number of errors : " << n_errors << std::endl;
-
 }
 
 int main(int argc, char** argv) {

examples/lgc/CMakeLists.txt

@@ -0,0 +1,21 @@
+# begin /* Set the application name. */
+set(APPLICATION_NAME lgc)
+# end /* Set the application name. */
+
+# begin /* Add CUDA executables */
+add_executable(${APPLICATION_NAME})
+
+set(SOURCE_LIST 
+    ${APPLICATION_NAME}.cu
+)
+
+target_sources(${APPLICATION_NAME} PRIVATE ${SOURCE_LIST})
+target_link_libraries(${APPLICATION_NAME} PRIVATE essentials)
+get_target_property(ESSENTIALS_ARCHITECTURES essentials CUDA_ARCHITECTURES)
+set_target_properties(${APPLICATION_NAME} 
+    PROPERTIES 
+        CUDA_ARCHITECTURES ${ESSENTIALS_ARCHITECTURES}
+) # XXX: Find a better way to inherit essentials properties.
+
+message("-- Example Added: ${APPLICATION_NAME}")
+# end /* Add CUDA executables */

examples/lgc/lgc.cu

@@ -0,0 +1,76 @@
+#include <gunrock/applications/lgc.hxx>
+
+using namespace gunrock;
+using namespace memory;
+
+void test_lgc(int num_arguments, char** argument_array) {
+  if (num_arguments != 2) {
+    std::cerr << "usage: ./bin/<program-name> filename.mtx" << std::endl;
+    exit(1);
+  }
+
+  // --
+  // Define types
+
+  using vertex_t = int;
+  using edge_t = int;
+  using weight_t = float;
+
+  using csr_t =
+      format::csr_t<memory_space_t::device, vertex_t, edge_t, weight_t>;
+  csr_t csr;
+
+  // --
+  // IO
+
+  std::string filename = argument_array[1];
+
+  if (util::is_market(filename)) {
+    io::matrix_market_t<vertex_t, edge_t, weight_t> mm;
+    csr.from_coo(mm.load(filename));
+  } else if (util::is_binary_csr(filename)) {
+    csr.read_binary(filename);
+  } else {
+    std::cerr << "Unknown file format: " << filename << std::endl;
+    exit(1);
+  }
+
+  // --
+  // Build graph
+
+  auto G = graph::build::from_csr<memory_space_t::device, graph::view_t::csr>(
+      csr.number_of_rows,               // rows
+      csr.number_of_columns,            // columns
+      csr.number_of_nonzeros,           // nonzeros
+      csr.row_offsets.data().get(),     // row_offsets
+      csr.column_indices.data().get(),  // column_indices
+      csr.nonzero_values.data().get()   // values
+  );  // supports row_indices and column_offsets (default = nullptr)
+
+  // --
+  // Params and memory allocation
+
+  srand(time(NULL));
+
+  weight_t alpha = 0.85;
+  weight_t tol = 1e-6;
+
+  vertex_t n_vertices = G.get_number_of_vertices();
+  thrust::device_vector<weight_t> p(n_vertices);
+
+  // --
+  // GPU Run
+
+  float gpu_elapsed = gunrock::pr::run(G, alpha, tol, p.data().get());
+
+  // --
+  // Log + Validate
+
+  std::cout << "GPU p[:40] = ";
+  gunrock::print::head<weight_t>(p, 40);
+  std::cout << "GPU Elapsed Time : " << gpu_elapsed << " (ms)" << std::endl;
+}
+
+int main(int argc, char** argv) {
+  test_lgc(argc, argv);
+}

examples/ppr/CMakeLists.txt

@@ -0,0 +1,21 @@
+# begin /* Set the application name. */
+set(APPLICATION_NAME ppr)
+# end /* Set the application name. */
+
+# begin /* Add CUDA executables */
+add_executable(${APPLICATION_NAME})
+
+set(SOURCE_LIST 
+    ${APPLICATION_NAME}.cu
+)
+
+target_sources(${APPLICATION_NAME} PRIVATE ${SOURCE_LIST})
+target_link_libraries(${APPLICATION_NAME} PRIVATE essentials)
+get_target_property(ESSENTIALS_ARCHITECTURES essentials CUDA_ARCHITECTURES)
+set_target_properties(${APPLICATION_NAME} 
+    PROPERTIES 
+        CUDA_ARCHITECTURES ${ESSENTIALS_ARCHITECTURES}
+) # XXX: Find a better way to inherit essentials properties.
+
+message("-- Example Added: ${APPLICATION_NAME}")
+# end /* Add CUDA executables */