diff --git a/cpp/include/cugraph/mtmg/detail/device_shared_wrapper.hpp b/cpp/include/cugraph/mtmg/detail/device_shared_wrapper.hpp index 3e4b2513a8d..5fbe7bc9f01 100644 --- a/cpp/include/cugraph/mtmg/detail/device_shared_wrapper.hpp +++ b/cpp/include/cugraph/mtmg/detail/device_shared_wrapper.hpp @@ -79,7 +79,6 @@ class device_shared_wrapper_t { objects_.insert(std::make_pair(local_rank, std::move(obj))); } - public: /** * @brief Get reference to an object for a particular thread * diff --git a/cpp/include/cugraph/mtmg/graph_view.hpp b/cpp/include/cugraph/mtmg/graph_view.hpp index 94347e016ea..8e202ab4904 100644 --- a/cpp/include/cugraph/mtmg/graph_view.hpp +++ b/cpp/include/cugraph/mtmg/graph_view.hpp @@ -27,8 +27,27 @@ namespace mtmg { * @brief Graph view for each GPU */ template -using graph_view_t = detail::device_shared_wrapper_t< - cugraph::graph_view_t>; +class graph_view_t : public detail::device_shared_wrapper_t< + cugraph::graph_view_t> { + public: + /** + * @brief Get the vertex_partition_view for this graph + */ + vertex_partition_view_t get_vertex_partition_view( + cugraph::mtmg::handle_t const& handle) const + { + return this->get(handle).local_vertex_partition_view(); + } + + /** + * @brief Get the vertex_partition_view for this graph + */ + std::vector get_vertex_partition_range_lasts( + cugraph::mtmg::handle_t const& handle) const + { + return this->get(handle).vertex_partition_range_lasts(); + } +}; } // namespace mtmg } // namespace cugraph diff --git a/cpp/include/cugraph/mtmg/vertex_result_view.hpp b/cpp/include/cugraph/mtmg/vertex_result_view.hpp index a349bb95333..42b80cea62f 100644 --- a/cpp/include/cugraph/mtmg/vertex_result_view.hpp +++ b/cpp/include/cugraph/mtmg/vertex_result_view.hpp @@ -39,11 +39,12 @@ class vertex_result_view_t : public detail::device_shared_device_span_t + template rmm::device_uvector gather( handle_t const& handle, raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, + std::vector const& vertex_partition_range_lasts, + cugraph::vertex_partition_view_t vertex_partition_view, std::optional>& renumber_map_view); }; diff --git a/cpp/src/mtmg/vertex_result.cu b/cpp/src/mtmg/vertex_result.cu index 9aad29d9507..5b1825656ff 100644 --- a/cpp/src/mtmg/vertex_result.cu +++ b/cpp/src/mtmg/vertex_result.cu @@ -27,15 +27,14 @@ namespace cugraph { namespace mtmg { template -template +template rmm::device_uvector vertex_result_view_t::gather( handle_t const& handle, raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, + std::vector const& vertex_partition_range_lasts, + vertex_partition_view_t vertex_partition_view, std::optional>& renumber_map_view) { - auto this_gpu_graph_view = graph_view.get(handle); - rmm::device_uvector local_vertices(vertices.size(), handle.get_stream()); rmm::device_uvector vertex_gpu_ids(vertices.size(), handle.get_stream()); rmm::device_uvector vertex_pos(vertices.size(), handle.get_stream()); @@ -47,11 +46,11 @@ rmm::device_uvector vertex_result_view_t::gather( cugraph::detail::sequence_fill( handle.get_stream(), vertex_pos.data(), vertex_pos.size(), size_t{0}); - rmm::device_uvector d_vertex_partition_range_lasts( - this_gpu_graph_view.vertex_partition_range_lasts().size(), handle.get_stream()); + rmm::device_uvector d_vertex_partition_range_lasts(vertex_partition_range_lasts.size(), + handle.get_stream()); raft::update_device(d_vertex_partition_range_lasts.data(), - this_gpu_graph_view.vertex_partition_range_lasts().data(), - this_gpu_graph_view.vertex_partition_range_lasts().size(), + vertex_partition_range_lasts.data(), + vertex_partition_range_lasts.size(), handle.get_stream()); if (renumber_map_view) { @@ -60,8 +59,8 @@ rmm::device_uvector vertex_result_view_t::gather( local_vertices.data(), local_vertices.size(), renumber_map_view->get(handle).data(), - this_gpu_graph_view.local_vertex_partition_range_first(), - this_gpu_graph_view.local_vertex_partition_range_last()); + vertex_partition_view.local_vertex_partition_range_first(), + vertex_partition_view.local_vertex_partition_range_last()); } auto const major_comm_size = @@ -89,8 +88,8 @@ rmm::device_uvector vertex_result_view_t::gather( auto& wrapped = this->get(handle); - auto vertex_partition = vertex_partition_device_view_t( - this_gpu_graph_view.local_vertex_partition_view()); + auto vertex_partition = + vertex_partition_device_view_t(vertex_partition_view); auto iter = thrust::make_transform_iterator(local_vertices.begin(), [vertex_partition] __device__(auto v) { @@ -130,145 +129,85 @@ rmm::device_uvector vertex_result_view_t::gather( template rmm::device_uvector vertex_result_view_t::gather( handle_t const& handle, raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, - std::optional>& renumber_map_view); - -template rmm::device_uvector vertex_result_view_t::gather( - handle_t const& handle, - raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, + std::vector const& vertex_partition_range_lasts, + vertex_partition_view_t vertex_partition_view, std::optional>& renumber_map_view); template rmm::device_uvector vertex_result_view_t::gather( handle_t const& handle, raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, + std::vector const& vertex_partition_range_lasts, + vertex_partition_view_t vertex_partition_view, std::optional>& renumber_map_view); template rmm::device_uvector vertex_result_view_t::gather( handle_t const& handle, raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, - std::optional>& renumber_map_view); - -template rmm::device_uvector vertex_result_view_t::gather( - handle_t const& handle, - raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, + std::vector const& vertex_partition_range_lasts, + vertex_partition_view_t vertex_partition_view, std::optional>& renumber_map_view); template rmm::device_uvector vertex_result_view_t::gather( handle_t const& handle, raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, + std::vector const& vertex_partition_range_lasts, + vertex_partition_view_t vertex_partition_view, std::optional>& renumber_map_view); -template rmm::device_uvector vertex_result_view_t::gather( +template rmm::device_uvector vertex_result_view_t::gather( handle_t const& handle, raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, + std::vector const& vertex_partition_range_lasts, + vertex_partition_view_t vertex_partition_view, std::optional>& renumber_map_view); -template rmm::device_uvector vertex_result_view_t::gather( - handle_t const& handle, - raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, - std::optional>& renumber_map_view); - -template rmm::device_uvector vertex_result_view_t::gather( +template rmm::device_uvector vertex_result_view_t::gather( handle_t const& handle, raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, + std::vector const& vertex_partition_range_lasts, + vertex_partition_view_t vertex_partition_view, std::optional>& renumber_map_view); -template rmm::device_uvector vertex_result_view_t::gather( - handle_t const& handle, - raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, - std::optional>& renumber_map_view); - -template rmm::device_uvector vertex_result_view_t::gather( +template rmm::device_uvector vertex_result_view_t::gather( handle_t const& handle, raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, + std::vector const& vertex_partition_range_lasts, + vertex_partition_view_t vertex_partition_view, std::optional>& renumber_map_view); -template rmm::device_uvector vertex_result_view_t::gather( +template rmm::device_uvector vertex_result_view_t::gather( handle_t const& handle, raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, + std::vector const& vertex_partition_range_lasts, + vertex_partition_view_t vertex_partition_view, std::optional>& renumber_map_view); template rmm::device_uvector vertex_result_view_t::gather( handle_t const& handle, raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, + std::vector const& vertex_partition_range_lasts, + vertex_partition_view_t vertex_partition_view, std::optional>& renumber_map_view); template rmm::device_uvector vertex_result_view_t::gather( handle_t const& handle, raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, + std::vector const& vertex_partition_range_lasts, + vertex_partition_view_t vertex_partition_view, std::optional>& renumber_map_view); template rmm::device_uvector vertex_result_view_t::gather( handle_t const& handle, raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, + std::vector const& vertex_partition_range_lasts, + vertex_partition_view_t vertex_partition_view, std::optional>& renumber_map_view); -template rmm::device_uvector vertex_result_view_t::gather( - handle_t const& handle, - raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, - std::optional>& renumber_map_view); - -template rmm::device_uvector vertex_result_view_t::gather( - handle_t const& handle, - raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, - std::optional>& renumber_map_view); - -template rmm::device_uvector vertex_result_view_t::gather( - handle_t const& handle, - raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, - std::optional>& renumber_map_view); - -template rmm::device_uvector vertex_result_view_t::gather( - handle_t const& handle, - raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, - std::optional>& renumber_map_view); - -template rmm::device_uvector vertex_result_view_t::gather( - handle_t const& handle, - raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, - std::optional>& renumber_map_view); - -template rmm::device_uvector vertex_result_view_t::gather( - handle_t const& handle, - raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, - std::optional>& renumber_map_view); - -template rmm::device_uvector vertex_result_view_t::gather( - handle_t const& handle, - raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, - std::optional>& renumber_map_view); - -template rmm::device_uvector vertex_result_view_t::gather( - handle_t const& handle, - raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, - std::optional>& renumber_map_view); - template rmm::device_uvector vertex_result_view_t::gather( handle_t const& handle, raft::device_span vertices, - cugraph::mtmg::graph_view_t const& graph_view, + std::vector const& vertex_partition_range_lasts, + vertex_partition_view_t vertex_partition_view, std::optional>& renumber_map_view); } // namespace mtmg diff --git a/cpp/tests/mtmg/multi_node_threaded_test.cu b/cpp/tests/mtmg/multi_node_threaded_test.cu index e5a7de07781..17aed4fdecf 100644 --- a/cpp/tests/mtmg/multi_node_threaded_test.cu +++ b/cpp/tests/mtmg/multi_node_threaded_test.cu @@ -311,7 +311,8 @@ class Tests_Multithreaded auto d_my_pageranks = pageranks_view.gather( thread_handle, raft::device_span{d_my_vertex_list.data(), d_my_vertex_list.size()}, - graph_view, + graph_view.get_vertex_partition_range_lasts(thread_handle), + graph_view.get_vertex_partition_view(thread_handle), renumber_map_view); std::vector my_pageranks(d_my_pageranks.size()); diff --git a/cpp/tests/mtmg/threaded_test.cu b/cpp/tests/mtmg/threaded_test.cu index 1a6a17eaa18..a5df0199cac 100644 --- a/cpp/tests/mtmg/threaded_test.cu +++ b/cpp/tests/mtmg/threaded_test.cu @@ -327,7 +327,8 @@ class Tests_Multithreaded auto d_my_pageranks = pageranks_view.gather( thread_handle, raft::device_span{d_my_vertex_list.data(), d_my_vertex_list.size()}, - graph_view, + graph_view.get_vertex_partition_range_lasts(thread_handle), + graph_view.get_vertex_partition_view(thread_handle), renumber_map_view); std::vector my_pageranks(d_my_pageranks.size()); diff --git a/cpp/tests/mtmg/threaded_test_louvain.cu b/cpp/tests/mtmg/threaded_test_louvain.cu index 3c39cc07811..c1395037646 100644 --- a/cpp/tests/mtmg/threaded_test_louvain.cu +++ b/cpp/tests/mtmg/threaded_test_louvain.cu @@ -207,8 +207,6 @@ class Tests_Multithreaded running_threads.resize(0); instance_manager->reset_threads(); - std::cout << "calling create graph" << std::endl; - for (int i = 0; i < num_gpus; ++i) { running_threads.emplace_back([&instance_manager, &graph, @@ -234,8 +232,6 @@ class Tests_Multithreaded edgelist.finalize_buffer(thread_handle); edgelist.consolidate_and_shuffle(thread_handle, false); - std::cout << "calling create_graph_from_edgelist" << std::endl; - cugraph::mtmg:: create_graph_from_edgelist( thread_handle, @@ -256,9 +252,6 @@ class Tests_Multithreaded running_threads.resize(0); instance_manager->reset_threads(); - std::cout << "created graph" << std::endl; - std::cout << " Has edge weights: " << (edge_weights ? "TRUE" : "FALSE") << std::endl; - graph_view = graph.view(); auto renumber_map_view = renumber_map ? std::make_optional(renumber_map->view()) : std::nullopt; @@ -291,13 +284,6 @@ class Tests_Multithreaded threshold, resolution); - sleep(thread_handle.get_rank()); - std::cout << "rank: " << thread_handle.get_rank() << std::endl; - raft::print_device_vector(" local_louvain_clusters", - local_louvain_clusters.data(), - local_louvain_clusters.size(), - std::cout); - louvain_clusters.set(thread_handle, std::move(local_louvain_clusters)); }); } @@ -307,8 +293,6 @@ class Tests_Multithreaded running_threads.resize(0); instance_manager->reset_threads(); - std::cout << "finished louvain" << std::endl; - std::vector, std::vector>> computed_clusters_v; std::mutex computed_clusters_lock{}; @@ -348,17 +332,11 @@ class Tests_Multithreaded my_vertex_list.size(), thread_handle.raft_handle().get_stream()); - sleep(thread_handle.get_rank()); - std::cout << "rank: " << thread_handle.get_rank() << std::endl; - raft::print_device_vector(" clusters", - louvain_clusters_view.get(thread_handle).data(), - louvain_clusters_view.get(thread_handle).size(), - std::cout); - auto d_my_clusters = louvain_clusters_view.gather( thread_handle, raft::device_span{d_my_vertex_list.data(), d_my_vertex_list.size()}, - graph_view, + graph_view.get_vertex_partition_range_lasts(thread_handle), + graph_view.get_vertex_partition_view(thread_handle), renumber_map_view); std::vector my_clusters(d_my_clusters.size()); @@ -380,8 +358,6 @@ class Tests_Multithreaded }); } - std::cout << "finished gather" << std::endl; - // Wait for CPU threads to complete std::for_each(running_threads.begin(), running_threads.end(), [](auto& t) { t.join(); }); running_threads.resize(0); @@ -441,17 +417,10 @@ class Tests_Multithreaded std::map h_cluster_map; std::map h_cluster_reverse_map; - raft::print_host_vector( - "h_sg_clusters", h_sg_clusters.data(), h_sg_clusters.size(), std::cout); - std::for_each( computed_clusters_v.begin(), computed_clusters_v.end(), [&h_sg_clusters, &h_cluster_map, &h_renumber_map, &h_cluster_reverse_map](auto t1) { - raft::print_host_vector( - " t1<0>", std::get<0>(t1).data(), std::get<0>(t1).size(), std::cout); - raft::print_host_vector( - " t1<1>", std::get<1>(t1).data(), std::get<1>(t1).size(), std::cout); std::for_each( thrust::make_zip_iterator(std::get<0>(t1).begin(), std::get<1>(t1).begin()), thrust::make_zip_iterator(std::get<0>(t1).end(), std::get<1>(t1).end()), @@ -462,15 +431,13 @@ class Tests_Multithreaded auto pos = std::find(h_renumber_map.begin(), h_renumber_map.end(), v); auto offset = std::distance(h_renumber_map.begin(), pos); - std::cout << " v = " << v << ", c = " << c << ", offset = " << offset << std::endl; - auto cluster_pos = h_cluster_map.find(c); if (cluster_pos == h_cluster_map.end()) { auto reverse_pos = h_cluster_reverse_map.find(h_sg_clusters[offset]); ASSERT_TRUE(reverse_pos != h_cluster_map.end()) << "two different cluster mappings"; - h_cluster_map.insert(std::make_pair(c, h_sg_clusters[v])); + h_cluster_map.insert(std::make_pair(c, h_sg_clusters[offset])); h_cluster_reverse_map.insert(std::make_pair(h_sg_clusters[offset], c)); } else { ASSERT_EQ(cluster_pos->second, h_sg_clusters[offset])