MNMG memory footprint improvement for low average vertex degree graph…

…s (part 2) (#1774)

This PR has the following 2 updates.

1. Update the type of the create_graph_from_edgelist functions's optional vertex list input parameter from std::tuple of a pointer and the array size to rmm::device_uvector r-value (to be consistent with the edge list input parameters and to allow releasing the memory earlier).
2. Avoid storing counts (vertex degrees) for isolated vertices (which implicitly have 0 degrees) to save memory.

Authors:
  - Seunghwa Kang (https://github.com/seunghwak)

Approvers:
  - Chuck Hastings (https://github.com/ChuckHastings)

URL: #1774

cpp/include/cugraph/graph_functions.hpp

@@ -347,8 +347,7 @@ extract_induced_subgraphs(
  * @param handle RAFT handle object to encapsulate resources (e.g. CUDA stream, communicator, and
  * handles to various CUDA libraries) to run graph algorithms.
  * @param vertex_span  If valid, part of the entire set of vertices in the graph to be renumbered.
- * The first tuple element is the pointer to the array and the second tuple element is the size of
- * the array. This parameter can be used to include isolated vertices. If multi-GPU, applying the
+ * This parameter can be used to include isolated vertices. If multi-GPU, applying the
  * compute_gpu_id_from_vertex_t to every vertex should return the local GPU ID for this function to
  * work (vertices should be pre-shuffled).
  * @param edgelist_rows Vector of edge row (source) vertex IDs.
@@ -369,7 +368,7 @@ template <typename vertex_t,
 std::tuple<cugraph::graph_t<vertex_t, edge_t, weight_t, store_transposed, multi_gpu>,
            std::optional<rmm::device_uvector<vertex_t>>>
 create_graph_from_edgelist(raft::handle_t const& handle,
-                           std::optional<std::tuple<vertex_t const*, vertex_t>> vertex_span,
+                           std::optional<rmm::device_uvector<vertex_t>>&& vertex_span,
                            rmm::device_uvector<vertex_t>&& edgelist_rows,
                            rmm::device_uvector<vertex_t>&& edgelist_cols,
                            std::optional<rmm::device_uvector<weight_t>>&& edgelist_weights,

cpp/src/structure/renumber_edgelist.cu

@@ -276,10 +276,13 @@ std::tuple<rmm::device_uvector<vertex_t>, std::vector<vertex_t>> compute_renumbe
   labels.shrink_to_fit(handle.get_stream());
   counts.shrink_to_fit(handle.get_stream());
 
-  // 4. if vertices != nullptr, add isolated vertices
+  auto num_non_isolated_vertices = static_cast<vertex_t>(labels.size());
+
+  // 4. if vertex_span.has_value() == true, append isolated vertices
 
-  rmm::device_uvector<vertex_t> isolated_vertices(0, handle.get_stream());
   if (vertex_span) {
+    rmm::device_uvector<vertex_t> isolated_vertices(0, handle.get_stream());
+
     auto [vertices, num_vertices] = *vertex_span;
     auto num_isolated_vertices    = thrust::count_if(
       rmm::exec_policy(handle.get_stream_view()),
@@ -296,30 +299,25 @@ std::tuple<rmm::device_uvector<vertex_t>, std::vector<vertex_t>> compute_renumbe
                     [label_first = labels.begin(), label_last = labels.end()] __device__(auto v) {
                       return !thrust::binary_search(thrust::seq, label_first, label_last, v);
                     });
-  }
 
-  if (isolated_vertices.size() > 0) {
-    labels.resize(labels.size() + isolated_vertices.size(), handle.get_stream());
-    counts.resize(labels.size(), handle.get_stream());
-    thrust::copy(rmm::exec_policy(handle.get_stream_view()),
-                 isolated_vertices.begin(),
-                 isolated_vertices.end(),
-                 labels.end() - isolated_vertices.size());
-    thrust::fill(rmm::exec_policy(handle.get_stream_view()),
-                 counts.end() - isolated_vertices.size(),
-                 counts.end(),
-                 edge_t{0});
+    if (isolated_vertices.size() > 0) {
+      labels.resize(labels.size() + isolated_vertices.size(), handle.get_stream());
+      thrust::copy(rmm::exec_policy(handle.get_stream_view()),
+                   isolated_vertices.begin(),
+                   isolated_vertices.end(),
+                   labels.end() - isolated_vertices.size());
+    }
   }
 
-  // 6. sort by degree
+  // 5. sort non-isolated vertices by degree
 
   thrust::sort_by_key(rmm::exec_policy(handle.get_stream_view()),
                       counts.begin(),
-                      counts.end(),
+                      counts.begin() + num_non_isolated_vertices,
                       labels.begin(),
                       thrust::greater<edge_t>());
 
-  // 7. compute segment_offsets
+  // 6. compute segment_offsets
 
   static_assert(detail::num_sparse_segments_per_vertex_partition == 3);
   static_assert((detail::low_degree_threshold <= detail::mid_degree_threshold) &&
@@ -353,7 +351,7 @@ std::tuple<rmm::device_uvector<vertex_t>, std::vector<vertex_t>> compute_renumbe
                                                   handle.get_stream());
 
   auto zero_vertex  = vertex_t{0};
-  auto vertex_count = static_cast<vertex_t>(counts.size());
+  auto vertex_count = static_cast<vertex_t>(labels.size());
   d_segment_offsets.set_element_async(0, zero_vertex, handle.get_stream());
   d_segment_offsets.set_element_async(
     num_segments_per_vertex_partition, vertex_count, handle.get_stream());
@@ -668,6 +666,7 @@ renumber_edgelist(
                                                               edgelist_const_major_vertices,
                                                               edgelist_const_minor_vertices,
                                                               edgelist_edge_counts);
+
   // 2. initialize partition_t object, number_of_vertices, and number_of_edges for the coarsened
   // graph
 

cpp/tests/community/mg_louvain_test.cpp

@@ -129,8 +129,7 @@ class Louvain_MG_Testfixture : public ::testing::TestWithParam<Louvain_Usecase>
       std::tie(*sg_graph, std::ignore) =
         cugraph::create_graph_from_edgelist<vertex_t, edge_t, weight_t, false, false>(
           handle,
-          std::optional<std::tuple<vertex_t const*, vertex_t>>{
-            std::make_tuple(d_vertices.data(), static_cast<vertex_t>(d_vertices.size()))},
+          std::move(d_vertices),
           std::move(d_edgelist_rows),
           std::move(d_edgelist_cols),
           std::move(d_edgelist_weights),

cpp/tests/components/wcc_graphs.cu

@@ -71,8 +71,7 @@ LineGraph_Usecase::construct_graph(raft::handle_t const& handle,
   return cugraph::
     create_graph_from_edgelist<vertex_t, edge_t, weight_t, store_transposed, multi_gpu>(
       handle,
-      std::optional<std::tuple<vertex_t const*, vertex_t>>{
-        std::make_tuple(vertices_v.data(), static_cast<vertex_t>(vertices_v.size()))},
+      std::move(vertices_v),
       std::move(src_v),
       std::move(dst_v),
       std::nullopt,

cpp/tests/utilities/matrix_market_file_utilities.cu

@@ -408,8 +408,7 @@ read_graph_from_matrix_market_file(raft::handle_t const& handle,
   return cugraph::
     create_graph_from_edgelist<vertex_t, edge_t, weight_t, store_transposed, multi_gpu>(
       handle,
-      std::optional<std::tuple<vertex_t const*, vertex_t>>{
-        std::make_tuple(d_vertices.data(), static_cast<vertex_t>(d_vertices.size()))},
+      std::move(d_vertices),
       std::move(d_edgelist_rows),
       std::move(d_edgelist_cols),
       std::move(d_edgelist_weights),

cpp/tests/utilities/rmat_utilities.cpp

@@ -184,8 +184,7 @@ generate_graph_from_rmat_params(raft::handle_t const& handle,
   return cugraph::
     create_graph_from_edgelist<vertex_t, edge_t, weight_t, store_transposed, multi_gpu>(
       handle,
-      std::optional<std::tuple<vertex_t const*, vertex_t>>{
-        std::make_tuple(d_vertices.data(), static_cast<vertex_t>(d_vertices.size()))},
+      std::move(d_vertices),
       std::move(d_edgelist_rows),
       std::move(d_edgelist_cols),
       std::move(d_edgelist_weights),