diff --git a/cpp/CMakeLists.txt b/cpp/CMakeLists.txt index 8ed460b8e17..956488a5f37 100644 --- a/cpp/CMakeLists.txt +++ b/cpp/CMakeLists.txt @@ -242,6 +242,8 @@ set(CUGRAPH_SOURCES src/centrality/katz_centrality_mg.cu src/centrality/eigenvector_centrality_sg.cu src/centrality/eigenvector_centrality_mg.cu + src/centrality/betweenness_centrality_sg.cu + src/centrality/betweenness_centrality_mg.cu src/tree/legacy/mst.cu src/components/weakly_connected_components_sg.cu src/components/weakly_connected_components_mg.cu @@ -347,6 +349,7 @@ add_library(cugraph_c src/c_api/katz.cpp src/c_api/centrality_result.cpp src/c_api/eigenvector_centrality.cpp + src/c_api/betweenness_centrality.cpp src/c_api/core_number.cpp src/c_api/core_result.cpp src/c_api/k_core.cpp diff --git a/cpp/include/cugraph/algorithms.hpp b/cpp/include/cugraph/algorithms.hpp index 8a5a19e0087..bf02887dc04 100644 --- a/cpp/include/cugraph/algorithms.hpp +++ b/cpp/include/cugraph/algorithms.hpp @@ -32,7 +32,9 @@ #include #include +#include #include +#include /** @ingroup cpp_api * @{ @@ -303,6 +305,92 @@ void edge_betweenness_centrality(const raft::handle_t& handle, vertex_t k = 0, vertex_t const* vertices = nullptr); +/** + * @brief Compute betweenness centrality for a graph + * + * Betweenness centrality for a vertex is the sum of the fraction of + * all pairs shortest paths that pass through the vertex. + * + * The current implementation does not support a weighted graph. + * + * If @p vertices is an optional variant. If it is not specified the algorithm + * will compute exact betweenness (compute betweenness using a traversal from all vertices). + * + * If @p vertices is specified as a vertex_t, it will compute approximate betweenness by + * random sampling @p vertices as the seeds of the traversals. + * + * If @p vertices is specified as a device_span, it will compute approximate betweenness + * using the provided @p vertices as the seeds of the traversals. + * + * @throws cugraph::logic_error when an error occurs. + * + * @tparam vertex_t Type of vertex identifiers. Needs to be an integral type. + * @tparam edge_t Type of edge identifiers. Needs to be an integral type. + * @tparam weight_t Type of edge weights. Needs to be a floating point type. + * @tparam multi_gpu Flag indicating whether template instantiation should target single-GPU (false) + * + * @param handle RAFT handle object to encapsulate resources (e.g. CUDA stream, communicator, and + * handles to various CUDA libraries) to run graph algorithms. + * @param graph_view Graph view object. + * @param vertices Optional, if specified this provides either a vertex_t count of how many + * random seeds to select, or a device_span identifying a list of pre-selected vertices + * to use as seeds for the traversals for approximating betweenness. + * @param normalized A flag indicating results should be normalized + * @param include_endpoints A flag indicating whether endpoints of a path should be counted + * @param do_expensive_check A flag to run expensive checks for input arguments (if set to `true`). + * + * @return device vector containing the centralities. + */ +template +rmm::device_uvector betweenness_centrality( + const raft::handle_t& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized = true, + bool const include_endpoints = false, + bool const do_expensive_check = false); + +/** + * @brief Compute edge betweenness centrality for a graph + * + * Betweenness centrality of an edge is the sum of the fraction of all-pairs shortest paths that + * pass through this edge. The weight parameter is currenlty not supported + * + * If @p vertices is an optional variant. If it is not specified the algorithm + * will compute exact betweenness (compute betweenness using a traversal from all vertices). + * + * If @p vertices is specified as a vertex_t, it will compute approximate betweenness by + * random sampling @p vertices as the seeds of the traversals. + * + * If @p vertices is specified as a device_span, it will compute approximate betweenness + * using the provided @p vertices as the seeds of the traversals. + * + * @throws cugraph::logic_error when an error occurs. + * + * @tparam vertex_t Type of vertex identifiers. Needs to be an integral type. + * @tparam edge_t Type of edge identifiers. Needs to be an integral type. + * @tparam weight_t Type of edge weights. Needs to be a floating point type. + * @tparam multi_gpu Flag indicating whether template instantiation should target single-GPU (false) + * + * @param handle RAFT handle object to encapsulate resources (e.g. CUDA stream, communicator, and + * handles to various CUDA libraries) to run graph algorithms. + * @param graph_view Graph view object. + * @param vertices Optional, if specified this provides either a vertex_t count of how many + * random seeds to select, or a device_span identifying a list of pre-selected vertices + * to use as seeds for the traversals for approximating betweenness. + * @param normalized A flag indicating whether or not to normalize the result + * @param do_expensive_check A flag to run expensive checks for input arguments (if set to `true`). + * + * @return device vector containing the centralities. + */ +template +rmm::device_uvector edge_betweenness_centrality( + const raft::handle_t& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool normalized = true, + bool do_expensive_check = false); + enum class cugraph_cc_t { CUGRAPH_WEAK = 0, ///> Weakly Connected Components CUGRAPH_STRONG, ///> Strongly Connected Components diff --git a/cpp/include/cugraph_c/centrality_algorithms.h b/cpp/include/cugraph_c/centrality_algorithms.h index e197ac4b403..27d15d0eb94 100644 --- a/cpp/include/cugraph_c/centrality_algorithms.h +++ b/cpp/include/cugraph_c/centrality_algorithms.h @@ -232,6 +232,128 @@ cugraph_error_code_t cugraph_katz_centrality(const cugraph_resource_handle_t* ha cugraph_centrality_result_t** result, cugraph_error_t** error); +/** + * @brief Compute betweenness centrality + * + * Betweenness can be computed exactly by specifying num_vertices as 0 and + * vertex_list as NULL. This will compute betweenness centrality by doing a + * traversal from every vertex and counting the frequency that a vertex appears on + * a shortest path. + * + * Approximate betweenness can be computed either by specifying num_vertices > 0, + * which will randomly pick the specified number of seeds; or by specifying a + * list of vertices that should be used as seeds for the BFS. + * + * Specifying both num_vertices > 0 and vertex_list as non-null will result in an + * error as the request is ambiguous. + * + * @param [in] handle Handle for accessing resources + * @param [in] graph Pointer to graph + * @param [in] num_vertices Number of vertices to randomly sample + * @param [in] vertex_list Optionally specify a device array containing a list of vertices + * to use as seeds for BFS + * @param [in] normalized Normalize + * @param [in] include_endpoints The traditional formulation of betweenness centrality does not + * include endpoints when considering a vertex to be on a shortest + * path. Setting this to true will consider the endpoints of a + * path to be part of the path. + * @param [in] do_expensive_check A flag to run expensive checks for input arguments (if set to + * `true`). + * @param [out] result Opaque pointer to betweenness centrality results + * @param [out] error Pointer to an error object storing details of any error. Will + * be populated if error code is not CUGRAPH_SUCCESS + * @return error code + */ +cugraph_error_code_t cugraph_betweenness_centrality( + const cugraph_resource_handle_t* handle, + cugraph_graph_t* graph, + size_t num_vertices, + const cugraph_type_erased_device_array_view_t* vertex_list, + bool_t normalized, + bool_t include_endpoints, + bool_t do_expensive_check, + cugraph_centrality_result_t** result, + cugraph_error_t** error); + +/** + * @brief Opaque edge centrality result type + */ +typedef struct { + int32_t align_; +} cugraph_edge_centrality_result_t; + +/** + * @brief Get the src vertex ids from an edge centrality result + * + * @param [in] result The result from an edge centrality algorithm + * @return type erased array of src vertex ids + */ +cugraph_type_erased_device_array_view_t* cugraph_edge_centrality_result_get_src_vertices( + cugraph_edge_centrality_result_t* result); + +/** + * @brief Get the dst vertex ids from an edge centrality result + * + * @param [in] result The result from an edge centrality algorithm + * @return type erased array of dst vertex ids + */ +cugraph_type_erased_device_array_view_t* cugraph_edge_centrality_result_get_dst_vertices( + cugraph_edge_centrality_result_t* result); + +/** + * @brief Get the centrality values from an edge centrality algorithm result + * + * @param [in] result The result from an edge centrality algorithm + * @return type erased array view of centrality values + */ +cugraph_type_erased_device_array_view_t* cugraph_edge_centrality_result_get_values( + cugraph_edge_centrality_result_t* result); + +/** + * @brief Free centrality result + * + * @param [in] result The result from a centrality algorithm + */ +void cugraph_edge_centrality_result_free(cugraph_edge_centrality_result_t* result); + +/** + * @brief Compute edge betweenness centrality + * + * Edge betweenness can be computed exactly by specifying num_vertices as 0 and + * vertex_list as NULL. This will compute betweenness centrality by doing a + * traversal from every vertex and counting the frequency that a edge appears on + * a shortest path. + * + * Approximate betweenness can be computed either by specifying num_vertices > 0, + * which will randomly pick the specified number of seeds; or by specifying a + * list of vertices that should be used as seeds for the BFS. + * + * Specifying both num_vertices > 0 and vertex_list as non-null will result in an + * error as the request is ambiguous. + * + * @param [in] handle Handle for accessing resources + * @param [in] graph Pointer to graph + * @param [in] num_vertices Number of vertices to randomly sample + * @param [in] vertex_list Optionally specify a device array containing a list of vertices + * to use as seeds for BFS + * @param [in] normalized Normalize + * @param [in] do_expensive_check A flag to run expensive checks for input arguments (if set to + * `true`). + * @param [out] result Opaque pointer to edge betweenness centrality results + * @param [out] error Pointer to an error object storing details of any error. Will + * be populated if error code is not CUGRAPH_SUCCESS + * @return error code + */ +cugraph_error_code_t cugraph_edge_betweenness_centrality( + const cugraph_resource_handle_t* handle, + cugraph_graph_t* graph, + size_t num_vertices, + const cugraph_type_erased_device_array_view_t* vertex_list, + bool_t normalized, + bool_t do_expensive_check, + cugraph_edge_centrality_result_t** result, + cugraph_error_t** error); + /** * @brief Opaque hits result type */ diff --git a/cpp/src/c_api/betweenness_centrality.cpp b/cpp/src/c_api/betweenness_centrality.cpp new file mode 100644 index 00000000000..96824ecee98 --- /dev/null +++ b/cpp/src/c_api/betweenness_centrality.cpp @@ -0,0 +1,240 @@ +/* + * Copyright (c) 2022, NVIDIA CORPORATION. + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governing permissions and + * limitations under the License. + */ + +#include + +#include +#include +#include +#include +#include + +#include +#include +#include + +#include + +namespace { + +struct betweenness_centrality_functor : public cugraph::c_api::abstract_functor { + raft::handle_t const& handle_; + cugraph::c_api::cugraph_graph_t* graph_{}; + size_t num_vertices_{}; + cugraph::c_api::cugraph_type_erased_device_array_view_t const* vertex_list_{}; + bool_t normalized_{}; + bool_t include_endpoints_{}; + bool do_expensive_check_{}; + cugraph::c_api::cugraph_centrality_result_t* result_{}; + + betweenness_centrality_functor(cugraph_resource_handle_t const* handle, + cugraph_graph_t* graph, + size_t num_vertices, + cugraph_type_erased_device_array_view_t const* vertex_list, + bool_t normalized, + bool_t include_endpoints, + bool do_expensive_check) + : abstract_functor(), + handle_(*reinterpret_cast(handle)->handle_), + graph_(reinterpret_cast(graph)), + num_vertices_(num_vertices), + vertex_list_(reinterpret_cast( + vertex_list)), + normalized_(normalized), + include_endpoints_(include_endpoints), + do_expensive_check_(do_expensive_check) + { + } + + template + void operator()() + { + // FIXME: Think about how to handle SG vice MG + if constexpr (!cugraph::is_candidate::value) { + unsupported(); + } else { + // Betweenness Centrality expects store_transposed == false + if constexpr (store_transposed) { + error_code_ = cugraph::c_api:: + transpose_storage( + handle_, graph_, error_.get()); + if (error_code_ != CUGRAPH_SUCCESS) return; + } + + auto graph = + reinterpret_cast*>( + graph_->graph_); + + auto graph_view = graph->view(); + + auto number_map = reinterpret_cast*>(graph_->number_map_); + + auto centralities = cugraph::betweenness_centrality( + handle_, + graph_view, + num_vertices_ > 0 + ? std::make_optional(std::variant>( + static_cast(num_vertices_))) + : vertex_list_ == NULL + ? std::optional>>{std::nullopt} + : std::make_optional>>( + raft::device_span{vertex_list_->as_type(), + vertex_list_->size_}), + normalized_, + include_endpoints_, + do_expensive_check_); + + rmm::device_uvector vertex_ids(graph_view.local_vertex_partition_range_size(), + handle_.get_stream()); + raft::copy(vertex_ids.data(), number_map->data(), vertex_ids.size(), handle_.get_stream()); + + result_ = new cugraph::c_api::cugraph_centrality_result_t{ + new cugraph::c_api::cugraph_type_erased_device_array_t(vertex_ids, graph_->vertex_type_), + new cugraph::c_api::cugraph_type_erased_device_array_t(centralities, graph_->weight_type_)}; + } + } +}; + +struct edge_betweenness_centrality_functor : public cugraph::c_api::abstract_functor { + raft::handle_t const& handle_; + cugraph::c_api::cugraph_graph_t* graph_{}; + size_t num_vertices_{}; + cugraph::c_api::cugraph_type_erased_device_array_view_t const* vertex_list_{}; + bool_t normalized_{}; + bool do_expensive_check_{}; + cugraph::c_api::cugraph_centrality_result_t* result_{}; + + edge_betweenness_centrality_functor(cugraph_resource_handle_t const* handle, + cugraph_graph_t* graph, + size_t num_vertices, + cugraph_type_erased_device_array_view_t const* vertex_list, + bool_t normalized, + bool do_expensive_check) + : abstract_functor(), + handle_(*reinterpret_cast(handle)->handle_), + graph_(reinterpret_cast(graph)), + num_vertices_(num_vertices), + vertex_list_(reinterpret_cast( + vertex_list)), + normalized_(normalized), + do_expensive_check_(do_expensive_check) + { + } + + template + void operator()() + { + // FIXME: Think about how to handle SG vice MG + if constexpr (!cugraph::is_candidate::value) { + unsupported(); + } else { + // Betweenness Centrality expects store_transposed == false + if constexpr (store_transposed) { + error_code_ = cugraph::c_api:: + transpose_storage( + handle_, graph_, error_.get()); + if (error_code_ != CUGRAPH_SUCCESS) return; + } + + auto graph = + reinterpret_cast*>( + graph_->graph_); + + auto graph_view = graph->view(); + + auto number_map = reinterpret_cast*>(graph_->number_map_); + + auto centralities = + cugraph::edge_betweenness_centrality( + handle_, + graph_view, + num_vertices_ > 0 + ? std::make_optional(std::variant>{ + static_cast(num_vertices_)}) + : vertex_list_ == NULL + ? std::optional>>{std::nullopt} + : std::make_optional>>( + raft::device_span{vertex_list_->as_type(), + vertex_list_->size_}), + normalized_, + do_expensive_check_); + + CUGRAPH_FAIL("Need to clean up return type"); + +#if 0 + result_ = new cugraph::c_api::cugraph_edge_centrality_result_t{ + new cugraph::c_api::cugraph_type_erased_device_array_t(src_ids, graph_->vertex_type_), + new cugraph::c_api::cugraph_type_erased_device_array_t(dst_ids, graph_->vertex_type_), + new cugraph::c_api::cugraph_type_erased_device_array_t(centralities, graph_->weight_type_)}; +#endif + } + } +}; + +} // namespace + +extern "C" cugraph_error_code_t cugraph_betweenness_centrality( + const cugraph_resource_handle_t* handle, + cugraph_graph_t* graph, + size_t num_vertices, + const cugraph_type_erased_device_array_view_t* vertex_list, + bool_t normalized, + bool_t include_endpoints, + bool_t do_expensive_check, + cugraph_centrality_result_t** result, + cugraph_error_t** error) +{ + CAPI_EXPECTS(((num_vertices == 0) || (vertex_list == NULL)), + CUGRAPH_INVALID_INPUT, + "Cannot specify both num_vertices and vertex_list", + *error); + + betweenness_centrality_functor functor( + handle, graph, num_vertices, vertex_list, normalized, include_endpoints, do_expensive_check); + + return cugraph::c_api::run_algorithm(graph, functor, result, error); +} + +extern "C" cugraph_error_code_t cugraph_edge_betweenness_centrality( + const cugraph_resource_handle_t* handle, + cugraph_graph_t* graph, + size_t num_vertices, + const cugraph_type_erased_device_array_view_t* vertex_list, + bool_t normalized, + bool_t do_expensive_check, + cugraph_edge_centrality_result_t** result, + cugraph_error_t** error) +{ + CAPI_EXPECTS(((num_vertices == 0) || (vertex_list == NULL)), + CUGRAPH_INVALID_INPUT, + "Cannot specify both num_vertices and vertex_list", + *error); + + edge_betweenness_centrality_functor functor( + handle, graph, num_vertices, vertex_list, normalized, do_expensive_check); + + return cugraph::c_api::run_algorithm(graph, functor, result, error); +} diff --git a/cpp/src/c_api/centrality_result.cpp b/cpp/src/c_api/centrality_result.cpp index 20794dff4bc..c3ded9fbd89 100644 --- a/cpp/src/c_api/centrality_result.cpp +++ b/cpp/src/c_api/centrality_result.cpp @@ -41,3 +41,40 @@ extern "C" void cugraph_centrality_result_free(cugraph_centrality_result_t* resu delete internal_pointer->values_; delete internal_pointer; } + +extern "C" cugraph_type_erased_device_array_view_t* cugraph_edge_centrality_result_get_src_vertices( + cugraph_edge_centrality_result_t* result) +{ + auto internal_pointer = + reinterpret_cast(result); + return reinterpret_cast( + internal_pointer->src_ids_->view()); +} + +extern "C" cugraph_type_erased_device_array_view_t* cugraph_edge_centrality_result_get_dst_vertices( + cugraph_edge_centrality_result_t* result) +{ + auto internal_pointer = + reinterpret_cast(result); + return reinterpret_cast( + internal_pointer->dst_ids_->view()); +} + +extern "C" cugraph_type_erased_device_array_view_t* cugraph_edge_centrality_result_get_values( + cugraph_edge_centrality_result_t* result) +{ + auto internal_pointer = + reinterpret_cast(result); + return reinterpret_cast( + internal_pointer->values_->view()); +} + +extern "C" void cugraph_edge_centrality_result_free(cugraph_edge_centrality_result_t* result) +{ + auto internal_pointer = + reinterpret_cast(result); + delete internal_pointer->src_ids_; + delete internal_pointer->dst_ids_; + delete internal_pointer->values_; + delete internal_pointer; +} diff --git a/cpp/src/c_api/centrality_result.hpp b/cpp/src/c_api/centrality_result.hpp index 230ad2f2ab9..e39db686152 100644 --- a/cpp/src/c_api/centrality_result.hpp +++ b/cpp/src/c_api/centrality_result.hpp @@ -26,5 +26,11 @@ struct cugraph_centrality_result_t { cugraph_type_erased_device_array_t* values_{}; }; +struct cugraph_edge_centrality_result_t { + cugraph_type_erased_device_array_t* src_ids_{}; + cugraph_type_erased_device_array_t* dst_ids_{}; + cugraph_type_erased_device_array_t* values_{}; +}; + } // namespace c_api } // namespace cugraph diff --git a/cpp/src/centrality/betweenness_centrality_impl.cuh b/cpp/src/centrality/betweenness_centrality_impl.cuh new file mode 100644 index 00000000000..8d1b0280365 --- /dev/null +++ b/cpp/src/centrality/betweenness_centrality_impl.cuh @@ -0,0 +1,83 @@ +/* + * Copyright (c) 2022, NVIDIA CORPORATION. + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governing permissions and + * limitations under the License. + */ +#pragma once + +#include +#include + +#include + +namespace cugraph { +namespace detail { + +template +rmm::device_uvector betweenness_centrality( + const raft::handle_t& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const include_endpoints, + bool const do_expensive_check) +{ + if (do_expensive_check) {} + + rmm::device_uvector centralities(graph_view.local_vertex_partition_range_size(), + handle.get_stream()); + + return centralities; +} + +template +rmm::device_uvector edge_betweenness_centrality( + const raft::handle_t& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const do_expensive_check) +{ + if (do_expensive_check) {} + + rmm::device_uvector centralities(0, handle.get_stream()); + + return centralities; +} + +} // namespace detail + +template +rmm::device_uvector betweenness_centrality( + const raft::handle_t& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const include_endpoints, + bool const do_expensive_check) +{ + CUGRAPH_FAIL("Not implemented"); +} + +template +rmm::device_uvector edge_betweenness_centrality( + const raft::handle_t& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const do_expensive_check) +{ + CUGRAPH_FAIL("Not implemented"); +} + +} // namespace cugraph diff --git a/cpp/src/centrality/betweenness_centrality_mg.cu b/cpp/src/centrality/betweenness_centrality_mg.cu new file mode 100644 index 00000000000..2238fd5b3cd --- /dev/null +++ b/cpp/src/centrality/betweenness_centrality_mg.cu @@ -0,0 +1,111 @@ +/* + * Copyright (c) 2022, NVIDIA CORPORATION. + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governing permissions and + * limitations under the License. + */ +#include + +namespace cugraph { + +// SG instantiation +template rmm::device_uvector betweenness_centrality( + raft::handle_t const& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const include_endpoints, + bool do_expensive_check); + +template rmm::device_uvector betweenness_centrality( + raft::handle_t const& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const include_endpoints, + bool do_expensive_check); + +template rmm::device_uvector betweenness_centrality( + raft::handle_t const& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const include_endpoints, + bool do_expensive_check); + +template rmm::device_uvector betweenness_centrality( + raft::handle_t const& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const include_endpoints, + bool do_expensive_check); + +template rmm::device_uvector betweenness_centrality( + raft::handle_t const& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const include_endpoints, + bool do_expensive_check); + +template rmm::device_uvector betweenness_centrality( + raft::handle_t const& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const include_endpoints, + bool do_expensive_check); + +template rmm::device_uvector edge_betweenness_centrality( + const raft::handle_t& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const do_expensive_check); + +template rmm::device_uvector edge_betweenness_centrality( + const raft::handle_t& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const do_expensive_check); + +template rmm::device_uvector edge_betweenness_centrality( + const raft::handle_t& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const do_expensive_check); + +template rmm::device_uvector edge_betweenness_centrality( + const raft::handle_t& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const do_expensive_check); + +template rmm::device_uvector edge_betweenness_centrality( + const raft::handle_t& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const do_expensive_check); + +template rmm::device_uvector edge_betweenness_centrality( + const raft::handle_t& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const do_expensive_check); + +} // namespace cugraph diff --git a/cpp/src/centrality/betweenness_centrality_sg.cu b/cpp/src/centrality/betweenness_centrality_sg.cu new file mode 100644 index 00000000000..78b768d4f50 --- /dev/null +++ b/cpp/src/centrality/betweenness_centrality_sg.cu @@ -0,0 +1,111 @@ +/* + * Copyright (c) 2022, NVIDIA CORPORATION. + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governing permissions and + * limitations under the License. + */ +#include + +namespace cugraph { + +// SG instantiation +template rmm::device_uvector betweenness_centrality( + raft::handle_t const& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const include_endpoints, + bool do_expensive_check); + +template rmm::device_uvector betweenness_centrality( + raft::handle_t const& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const include_endpoints, + bool do_expensive_check); + +template rmm::device_uvector betweenness_centrality( + raft::handle_t const& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const include_endpoints, + bool do_expensive_check); + +template rmm::device_uvector betweenness_centrality( + raft::handle_t const& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const include_endpoints, + bool do_expensive_check); + +template rmm::device_uvector betweenness_centrality( + raft::handle_t const& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const include_endpoints, + bool do_expensive_check); + +template rmm::device_uvector betweenness_centrality( + raft::handle_t const& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const include_endpoints, + bool do_expensive_check); + +template rmm::device_uvector edge_betweenness_centrality( + const raft::handle_t& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const do_expensive_check); + +template rmm::device_uvector edge_betweenness_centrality( + const raft::handle_t& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const do_expensive_check); + +template rmm::device_uvector edge_betweenness_centrality( + const raft::handle_t& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const do_expensive_check); + +template rmm::device_uvector edge_betweenness_centrality( + const raft::handle_t& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const do_expensive_check); + +template rmm::device_uvector edge_betweenness_centrality( + const raft::handle_t& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const do_expensive_check); + +template rmm::device_uvector edge_betweenness_centrality( + const raft::handle_t& handle, + graph_view_t const& graph_view, + std::optional>> vertices, + bool const normalized, + bool const do_expensive_check); + +} // namespace cugraph diff --git a/cpp/tests/CMakeLists.txt b/cpp/tests/CMakeLists.txt index 8ce84bff507..8f7649a0ed4 100644 --- a/cpp/tests/CMakeLists.txt +++ b/cpp/tests/CMakeLists.txt @@ -26,6 +26,7 @@ add_library(cugraphtestutil STATIC utilities/test_utilities_sg.cu utilities/test_utilities_mg.cu link_prediction/similarity_compare.cpp + centrality/betweenness_centrality_validate.cu structure/induced_subgraph_validate.cu components/wcc_graphs.cu sampling/random_walks_check_sg.cu @@ -331,6 +332,11 @@ ConfigureTest(KATZ_CENTRALITY_TEST centrality/katz_centrality_test.cpp) # - EIGENVECTOR_CENTRALITY tests ------------------------------------------------------------------------- ConfigureTest(EIGENVECTOR_CENTRALITY_TEST centrality/eigenvector_centrality_test.cpp) +################################################################################################### +# - BETWEENNESS_CENTRALITY tests ------------------------------------------------------------------------- +ConfigureTest(BETWEENNESS_CENTRALITY_TEST centrality/betweenness_centrality_test.cpp) +ConfigureTest(EDGE_BETWEENNESS_CENTRALITY_TEST centrality/edge_betweenness_centrality_test.cpp) + ################################################################################################### # - WEAKLY CONNECTED COMPONENTS tests ------------------------------------------------------------- ConfigureTest(WEAKLY_CONNECTED_COMPONENTS_TEST components/weakly_connected_components_test.cpp) @@ -443,9 +449,14 @@ if(BUILD_CUGRAPH_MG_TESTS) ConfigureTestMG(MG_KATZ_CENTRALITY_TEST centrality/mg_katz_centrality_test.cpp) ############################################################################################### - # - MG EIGENVECTOR CENTRALITY tests ----------------------------------------------------------- + # - MG EIGENVECTOR CENTRALITY tests ------------------------------------------------------------------ ConfigureTestMG(MG_EIGENVECTOR_CENTRALITY_TEST centrality/mg_eigenvector_centrality_test.cpp) + ############################################################################################### + # - MG BETWEENNESS CENTRALITY tests ----------------------------------------------------------- + ConfigureTestMG(MG_BETWEENNESS_CENTRALITY_TEST centrality/mg_betweenness_centrality_test.cpp) + ConfigureTestMG(MG_EDGE_BETWEENNESS_CENTRALITY_TEST centrality/mg_edge_betweenness_centrality_test.cpp) + ############################################################################################### # - MG BFS tests ------------------------------------------------------------------------------ ConfigureTestMG(MG_BFS_TEST traversal/mg_bfs_test.cpp) @@ -573,6 +584,7 @@ if(BUILD_CUGRAPH_MG_TESTS) ConfigureCTestMG(MG_CAPI_STRONGLY_CONNECTED_COMPONENTS_TEST c_api/mg_strongly_connected_components_test.c c_api/mg_test_utils.cpp) ConfigureCTestMG(MG_CAPI_KATZ_TEST c_api/mg_katz_test.c c_api/mg_test_utils.cpp) ConfigureCTestMG(MG_CAPI_EIGENVECTOR_CENTRALITY_TEST c_api/mg_eigenvector_centrality_test.c c_api/mg_test_utils.cpp) + ConfigureCTestMG(MG_CAPI_BETWEENNESS_CENTRALITY_TEST c_api/mg_betweenness_centrality_test.c c_api/mg_test_utils.cpp) ConfigureCTestMG(MG_CAPI_HITS_TEST c_api/mg_hits_test.c c_api/mg_test_utils.cpp) ConfigureCTestMG(MG_CAPI_UNIFORM_NEIGHBOR_SAMPLE_TEST c_api/mg_uniform_neighbor_sample_test.c c_api/mg_test_utils.cpp) ConfigureCTestMG(MG_CAPI_RANDOM_WALKS_TEST c_api/mg_random_walks_test.c c_api/mg_test_utils.cpp) @@ -619,6 +631,7 @@ ConfigureCTest(CAPI_CREATE_GRAPH_TEST c_api/create_graph_test.c) ConfigureCTest(CAPI_PAGERANK_TEST c_api/pagerank_test.c) ConfigureCTest(CAPI_KATZ_TEST c_api/katz_test.c) ConfigureCTest(CAPI_EIGENVECTOR_CENTRALITY_TEST c_api/eigenvector_centrality_test.c) +ConfigureCTest(CAPI_BETWEENNESS_CENTRALITY_TEST c_api/betweenness_centrality_test.c) ConfigureCTest(CAPI_HITS_TEST c_api/hits_test.c) ConfigureCTest(CAPI_BFS_TEST c_api/bfs_test.c) ConfigureCTest(CAPI_SSSP_TEST c_api/sssp_test.c) diff --git a/cpp/tests/c_api/betweenness_centrality_test.c b/cpp/tests/c_api/betweenness_centrality_test.c new file mode 100644 index 00000000000..5056fc09bc5 --- /dev/null +++ b/cpp/tests/c_api/betweenness_centrality_test.c @@ -0,0 +1,123 @@ +/* + * Copyright (c) 2022, NVIDIA CORPORATION. + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governing permissions and + * limitations under the License. + */ + +#include "c_test_utils.h" /* RUN_TEST */ + +#include +#include + +#include + +typedef int32_t vertex_t; +typedef int32_t edge_t; +typedef float weight_t; + +int generic_betweenness_centrality_test(vertex_t* h_src, + vertex_t* h_dst, + weight_t* h_wgt, + weight_t* h_result, + size_t num_vertices, + size_t num_edges, + bool_t store_transposed, + size_t num_vertices_to_sample) +{ + int test_ret_value = 0; + + cugraph_error_code_t ret_code = CUGRAPH_SUCCESS; + cugraph_error_t* ret_error; + + cugraph_resource_handle_t* p_handle = NULL; + cugraph_graph_t* p_graph = NULL; + cugraph_centrality_result_t* p_result = NULL; + + p_handle = cugraph_create_resource_handle(NULL); + TEST_ASSERT(test_ret_value, p_handle != NULL, "resource handle creation failed."); + + ret_code = create_test_graph( + p_handle, h_src, h_dst, h_wgt, num_edges, store_transposed, FALSE, FALSE, &p_graph, &ret_error); + + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "create_test_graph failed."); + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, cugraph_error_message(ret_error)); + + ret_code = cugraph_betweenness_centrality( + p_handle, p_graph, num_vertices_to_sample, NULL, FALSE, FALSE, FALSE, &p_result, &ret_error); +#if 1 + TEST_ASSERT(test_ret_value, ret_code != CUGRAPH_SUCCESS, "cugraph_betweenness_centrality should have failed"); +#else + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, cugraph_error_message(ret_error)); + TEST_ASSERT( + test_ret_value, ret_code == CUGRAPH_SUCCESS, "cugraph_betweenness_centrality failed."); + + cugraph_type_erased_device_array_view_t* vertices; + cugraph_type_erased_device_array_view_t* centralities; + + vertices = cugraph_centrality_result_get_vertices(p_result); + centralities = cugraph_centrality_result_get_values(p_result); + + vertex_t h_vertices[num_vertices]; + weight_t h_centralities[num_vertices]; + + ret_code = cugraph_type_erased_device_array_view_copy_to_host( + p_handle, (byte_t*)h_vertices, vertices, &ret_error); + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "copy_to_host failed."); + + ret_code = cugraph_type_erased_device_array_view_copy_to_host( + p_handle, (byte_t*)h_centralities, centralities, &ret_error); + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "copy_to_host failed."); + + for (int i = 0; (i < num_vertices) && (test_ret_value == 0); ++i) { + TEST_ASSERT(test_ret_value, + nearlyEqual(h_result[h_vertices[i]], h_centralities[i], 0.001), + "centralities results don't match"); + } + + cugraph_centrality_result_free(p_result); +#endif + + cugraph_sg_graph_free(p_graph); + cugraph_free_resource_handle(p_handle); + cugraph_error_free(ret_error); + + return test_ret_value; +} + +int test_betweenness_centrality() +{ + size_t num_edges = 16; + size_t num_vertices = 6; + + vertex_t h_src[] = {0, 1, 1, 2, 2, 2, 3, 4, 1, 3, 4, 0, 1, 3, 5, 5}; + vertex_t h_dst[] = {1, 3, 4, 0, 1, 3, 5, 5, 0, 1, 1, 2, 2, 2, 3, 4}; + weight_t h_wgt[] = { + 0.1f, 2.1f, 1.1f, 5.1f, 3.1f, 4.1f, 7.2f, 3.2f, 0.1f, 2.1f, 1.1f, 5.1f, 3.1f, 4.1f, 7.2f, 3.2f}; + weight_t h_result[] = {0.236325, 0.292055, 0.458457, 0.60533, 0.190498, 0.495942}; + + double epsilon = 1e-6; + size_t max_iterations = 200; + + // Betweenness centrality wants store_transposed = FAlSE + return generic_betweenness_centrality_test( + h_src, h_dst, h_wgt, h_result, num_vertices, num_edges, FALSE, 5); +} + +/******************************************************************************/ + +int main(int argc, char** argv) +{ + int result = 0; + result |= RUN_TEST(test_betweenness_centrality); + return result; +} diff --git a/cpp/tests/c_api/edge_betweenness_centrality.c b/cpp/tests/c_api/edge_betweenness_centrality.c new file mode 100644 index 00000000000..75dbe464654 --- /dev/null +++ b/cpp/tests/c_api/edge_betweenness_centrality.c @@ -0,0 +1,118 @@ +/* + * Copyright (c) 2022, NVIDIA CORPORATION. + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governing permissions and + * limitations under the License. + */ + +#include "c_test_utils.h" /* RUN_TEST */ + +#include +#include + +#include + +typedef int32_t vertex_t; +typedef int32_t edge_t; +typedef float weight_t; + +int generic_edge_betweenness_centrality_test(vertex_t* h_src, + vertex_t* h_dst, + weight_t* h_wgt, + weight_t* h_result, + size_t num_vertices, + size_t num_edges, + bool_t store_transposed, + size_t num_vertices_to_sample) +{ + int test_ret_value = 0; + + cugraph_error_code_t ret_code = CUGRAPH_SUCCESS; + cugraph_error_t* ret_error; + + cugraph_resource_handle_t* p_handle = NULL; + cugraph_graph_t* p_graph = NULL; + cugraph_centrality_result_t* p_result = NULL; + + p_handle = cugraph_create_resource_handle(NULL); + TEST_ASSERT(test_ret_value, p_handle != NULL, "resource handle creation failed."); + + ret_code = create_test_graph( + p_handle, h_src, h_dst, h_wgt, num_edges, store_transposed, FALSE, FALSE, &p_graph, &ret_error); + + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "create_test_graph failed."); + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, cugraph_error_message(ret_error)); + + ret_code = cugraph_edge_betweenness_centrality( + p_handle, p_graph, num_vertices_to_sample, NULL, FALSE, FALSE, &p_result, &ret_error); + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, cugraph_error_message(ret_error)); + TEST_ASSERT( + test_ret_value, ret_code == CUGRAPH_SUCCESS, "cugraph_edge_betweenness_centrality failed."); + + cugraph_type_erased_device_array_view_t* vertices; + cugraph_type_erased_device_array_view_t* centralities; + + vertices = cugraph_centrality_result_get_vertices(p_result); + centralities = cugraph_centrality_result_get_values(p_result); + + vertex_t h_vertices[num_vertices]; + weight_t h_centralities[num_vertices]; + + ret_code = cugraph_type_erased_device_array_view_copy_to_host( + p_handle, (byte_t*)h_vertices, vertices, &ret_error); + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "copy_to_host failed."); + + ret_code = cugraph_type_erased_device_array_view_copy_to_host( + p_handle, (byte_t*)h_centralities, centralities, &ret_error); + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "copy_to_host failed."); + + for (int i = 0; (i < num_vertices) && (test_ret_value == 0); ++i) { + TEST_ASSERT(test_ret_value, + nearlyEqual(h_result[h_vertices[i]], h_centralities[i], 0.001), + "centralities results don't match"); + } + + cugraph_centrality_result_free(p_result); + cugraph_sg_graph_free(p_graph); + cugraph_free_resource_handle(p_handle); + cugraph_error_free(ret_error); + + return test_ret_value; +} + +int test_edge_betweenness_centrality() +{ + size_t num_edges = 16; + size_t num_vertices = 6; + + vertex_t h_src[] = {0, 1, 1, 2, 2, 2, 3, 4, 1, 3, 4, 0, 1, 3, 5, 5}; + vertex_t h_dst[] = {1, 3, 4, 0, 1, 3, 5, 5, 0, 1, 1, 2, 2, 2, 3, 4}; + weight_t h_wgt[] = { + 0.1f, 2.1f, 1.1f, 5.1f, 3.1f, 4.1f, 7.2f, 3.2f, 0.1f, 2.1f, 1.1f, 5.1f, 3.1f, 4.1f, 7.2f, 3.2f}; + weight_t h_result[] = {0.236325, 0.292055, 0.458457, 0.60533, 0.190498, 0.495942}; + + double epsilon = 1e-6; + size_t max_iterations = 200; + + // Eigenvector centrality wants store_transposed = TRUE + return generic_edge_betweenness_centrality_test( + h_src, h_dst, h_wgt, h_result, num_vertices, num_edges, TRUE, 5); +} + +/******************************************************************************/ + +int main(int argc, char** argv) +{ + int result = 0; + result |= RUN_TEST(test_edge_betweenness_centrality); + return result; +} diff --git a/cpp/tests/c_api/mg_betweenness_centrality_test.c b/cpp/tests/c_api/mg_betweenness_centrality_test.c new file mode 100644 index 00000000000..c9a714c4bf4 --- /dev/null +++ b/cpp/tests/c_api/mg_betweenness_centrality_test.c @@ -0,0 +1,164 @@ +/* + * Copyright (c) 2022, NVIDIA CORPORATION. + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governing permissions and + * limitations under the License. + */ + +#include "mg_test_utils.h" /* RUN_TEST */ + +#include +#include + +#include + +typedef int32_t vertex_t; +typedef int32_t edge_t; +typedef float weight_t; + +int generic_betweenness_centrality_test(const cugraph_resource_handle_t* handle, + vertex_t* h_src, + vertex_t* h_dst, + weight_t* h_wgt, + weight_t* h_result, + size_t num_vertices, + size_t num_edges, + bool_t store_transposed, + size_t num_vertices_to_sample) +{ + int test_ret_value = 0; + + cugraph_error_code_t ret_code = CUGRAPH_SUCCESS; + cugraph_error_t* ret_error; + + cugraph_graph_t* p_graph = NULL; + cugraph_centrality_result_t* p_result = NULL; + + ret_code = create_mg_test_graph( + handle, h_src, h_dst, h_wgt, num_edges, store_transposed, FALSE, &p_graph, &ret_error); + + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "create_mg_test_graph failed."); + + ret_code = cugraph_betweenness_centrality( + handle, p_graph, num_vertices_to_sample, NULL, FALSE, FALSE, FALSE, &p_result, &ret_error); +#if 1 + TEST_ASSERT(test_ret_value, ret_code != CUGRAPH_SUCCESS, "cugraph_betweenness_centrality should have failed"); +#else + TEST_ASSERT( + test_ret_value, ret_code == CUGRAPH_SUCCESS, "cugraph_betweenness_centrality failed."); + + // NOTE: Because we get back vertex ids and centralities, we can simply compare + // the returned values with the expected results for the entire + // graph. Each GPU will have a subset of the total vertices, so + // they will do a subset of the comparisons. + cugraph_type_erased_device_array_view_t* vertices; + cugraph_type_erased_device_array_view_t* centralities; + + vertices = cugraph_centrality_result_get_vertices(p_result); + centralities = cugraph_centrality_result_get_values(p_result); + + vertex_t h_vertices[num_vertices]; + weight_t h_centralities[num_vertices]; + + ret_code = cugraph_type_erased_device_array_view_copy_to_host( + handle, (byte_t*)h_vertices, vertices, &ret_error); + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "copy_to_host failed."); + + ret_code = cugraph_type_erased_device_array_view_copy_to_host( + handle, (byte_t*)h_centralities, centralities, &ret_error); + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "copy_to_host failed."); + + size_t num_local_vertices = cugraph_type_erased_device_array_view_size(vertices); + + for (int i = 0; (i < num_local_vertices) && (test_ret_value == 0); ++i) { + TEST_ASSERT(test_ret_value, + nearlyEqual(h_result[h_vertices[i]], h_centralities[i], 0.001), + "betweenness centrality results don't match"); + } + + cugraph_centrality_result_free(p_result); +#endif + + cugraph_mg_graph_free(p_graph); + cugraph_error_free(ret_error); + + return test_ret_value; +} + +int test_betweenness_centrality(const cugraph_resource_handle_t* handle) +{ + size_t num_edges = 16; + size_t num_vertices = 6; + + vertex_t h_src[] = {0, 1, 1, 2, 2, 2, 3, 4, 1, 3, 4, 0, 1, 3, 5, 5}; + vertex_t h_dst[] = {1, 3, 4, 0, 1, 3, 5, 5, 0, 1, 1, 2, 2, 2, 3, 4}; + weight_t h_wgt[] = { + 0.1f, 2.1f, 1.1f, 5.1f, 3.1f, 4.1f, 7.2f, 3.2f, 0.1f, 2.1f, 1.1f, 5.1f, 3.1f, 4.1f, 7.2f, 3.2f}; + weight_t h_result[] = {0.236374, 0.292046, 0.458369, 0.605472, 0.190544, 0.495814}; + + double epsilon = 1e-6; + size_t max_iterations = 200; + + // Betweenness centrality wants store_transposed = FALSE + return generic_betweenness_centrality_test( + handle, h_src, h_dst, h_wgt, h_result, num_vertices, num_edges, FALSE, 5); +} + +/******************************************************************************/ + +int main(int argc, char** argv) +{ + // Set up MPI: + int comm_rank; + int comm_size; + int num_gpus_per_node; + cudaError_t status; + int mpi_status; + int result = 0; + cugraph_resource_handle_t* handle = NULL; + cugraph_error_t* ret_error; + cugraph_error_code_t ret_code = CUGRAPH_SUCCESS; + int prows = 1; + + C_MPI_TRY(MPI_Init(&argc, &argv)); + C_MPI_TRY(MPI_Comm_rank(MPI_COMM_WORLD, &comm_rank)); + C_MPI_TRY(MPI_Comm_size(MPI_COMM_WORLD, &comm_size)); + C_CUDA_TRY(cudaGetDeviceCount(&num_gpus_per_node)); + C_CUDA_TRY(cudaSetDevice(comm_rank % num_gpus_per_node)); + +#if 0 + // TODO: Need something a bit more sophisticated for bigger systems + prows = (int)sqrt((double)comm_size); + while (comm_size % prows != 0) { + --prows; + } + + ret_code = cugraph_resource_handle_init_comms(handle, prows, &ret_error); + TEST_ASSERT(result, ret_code == CUGRAPH_SUCCESS, "handle create failed."); + TEST_ASSERT(result, ret_code == CUGRAPH_SUCCESS, cugraph_error_message(ret_error)); +#endif + + void* raft_handle = create_raft_handle(prows); + handle = cugraph_create_resource_handle(raft_handle); + + if (result == 0) { + result |= RUN_MG_TEST(test_betweenness_centrality, handle); + + cugraph_free_resource_handle(handle); + } + + free_raft_handle(raft_handle); + + C_MPI_TRY(MPI_Finalize()); + + return result; +} diff --git a/cpp/tests/c_api/mg_edge_betweenness_centrality.c b/cpp/tests/c_api/mg_edge_betweenness_centrality.c new file mode 100644 index 00000000000..24e109f2f49 --- /dev/null +++ b/cpp/tests/c_api/mg_edge_betweenness_centrality.c @@ -0,0 +1,159 @@ +/* + * Copyright (c) 2022, NVIDIA CORPORATION. + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governing permissions and + * limitations under the License. + */ + +#include "mg_test_utils.h" /* RUN_TEST */ + +#include +#include + +#include + +typedef int32_t vertex_t; +typedef int32_t edge_t; +typedef float weight_t; + +int generic_edge_betweenness_centrality_test(const cugraph_resource_handle_t* handle, + vertex_t* h_src, + vertex_t* h_dst, + weight_t* h_wgt, + weight_t* h_result, + size_t num_vertices, + size_t num_edges, + bool_t store_transposed, + size_t num_vertices_to_sample) +{ + int test_ret_value = 0; + + cugraph_error_code_t ret_code = CUGRAPH_SUCCESS; + cugraph_error_t* ret_error; + + cugraph_graph_t* p_graph = NULL; + cugraph_centrality_result_t* p_result = NULL; + + ret_code = create_mg_test_graph( + handle, h_src, h_dst, h_wgt, num_edges, store_transposed, FALSE, &p_graph, &ret_error); + + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "create_mg_test_graph failed."); + + ret_code = cugraph_edge_betweenness_centrality( + handle, p_graph, num_vertices_to_sample, NULL, FALSE, FALSE, &p_result, &ret_error); + TEST_ASSERT( + test_ret_value, ret_code == CUGRAPH_SUCCESS, "cugraph_edge_betweenness_centrality failed."); + + // NOTE: Because we get back vertex ids and centralities, we can simply compare + // the returned values with the expected results for the entire + // graph. Each GPU will have a subset of the total vertices, so + // they will do a subset of the comparisons. + cugraph_type_erased_device_array_view_t* vertices; + cugraph_type_erased_device_array_view_t* centralities; + + vertices = cugraph_centrality_result_get_vertices(p_result); + centralities = cugraph_centrality_result_get_values(p_result); + + vertex_t h_vertices[num_vertices]; + weight_t h_centralities[num_vertices]; + + ret_code = cugraph_type_erased_device_array_view_copy_to_host( + handle, (byte_t*)h_vertices, vertices, &ret_error); + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "copy_to_host failed."); + + ret_code = cugraph_type_erased_device_array_view_copy_to_host( + handle, (byte_t*)h_centralities, centralities, &ret_error); + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "copy_to_host failed."); + + size_t num_local_vertices = cugraph_type_erased_device_array_view_size(vertices); + + for (int i = 0; (i < num_local_vertices) && (test_ret_value == 0); ++i) { + TEST_ASSERT(test_ret_value, + nearlyEqual(h_result[h_vertices[i]], h_centralities[i], 0.001), + "betweenness centrality results don't match"); + } + + cugraph_centrality_result_free(p_result); + cugraph_mg_graph_free(p_graph); + cugraph_error_free(ret_error); + + return test_ret_value; +} + +int test_edge_betweenness_centrality(const cugraph_resource_handle_t* handle) +{ + size_t num_edges = 16; + size_t num_vertices = 6; + + vertex_t h_src[] = {0, 1, 1, 2, 2, 2, 3, 4, 1, 3, 4, 0, 1, 3, 5, 5}; + vertex_t h_dst[] = {1, 3, 4, 0, 1, 3, 5, 5, 0, 1, 1, 2, 2, 2, 3, 4}; + weight_t h_wgt[] = { + 0.1f, 2.1f, 1.1f, 5.1f, 3.1f, 4.1f, 7.2f, 3.2f, 0.1f, 2.1f, 1.1f, 5.1f, 3.1f, 4.1f, 7.2f, 3.2f}; + weight_t h_result[] = {0.236374, 0.292046, 0.458369, 0.605472, 0.190544, 0.495814}; + + double epsilon = 1e-6; + size_t max_iterations = 200; + + // Eigenvector centrality wants store_transposed = TRUE + return generic_edge_betweenness_centrality_test( + handle, h_src, h_dst, h_wgt, h_result, num_vertices, num_edges, TRUE, 5); +} + +/******************************************************************************/ + +int main(int argc, char** argv) +{ + // Set up MPI: + int comm_rank; + int comm_size; + int num_gpus_per_node; + cudaError_t status; + int mpi_status; + int result = 0; + cugraph_resource_handle_t* handle = NULL; + cugraph_error_t* ret_error; + cugraph_error_code_t ret_code = CUGRAPH_SUCCESS; + int prows = 1; + + C_MPI_TRY(MPI_Init(&argc, &argv)); + C_MPI_TRY(MPI_Comm_rank(MPI_COMM_WORLD, &comm_rank)); + C_MPI_TRY(MPI_Comm_size(MPI_COMM_WORLD, &comm_size)); + C_CUDA_TRY(cudaGetDeviceCount(&num_gpus_per_node)); + C_CUDA_TRY(cudaSetDevice(comm_rank % num_gpus_per_node)); + +#if 0 + // TODO: Need something a bit more sophisticated for bigger systems + prows = (int)sqrt((double)comm_size); + while (comm_size % prows != 0) { + --prows; + } + + ret_code = cugraph_resource_handle_init_comms(handle, prows, &ret_error); + TEST_ASSERT(result, ret_code == CUGRAPH_SUCCESS, "handle create failed."); + TEST_ASSERT(result, ret_code == CUGRAPH_SUCCESS, cugraph_error_message(ret_error)); +#endif + + void* raft_handle = create_raft_handle(prows); + handle = cugraph_create_resource_handle(raft_handle); + + if (result == 0) { + result |= RUN_MG_TEST(test_edge_betweenness_centrality, handle); + + cugraph_free_resource_handle(handle); + } + + free_raft_handle(raft_handle); + + C_MPI_TRY(MPI_Finalize()); + + return result; +} diff --git a/cpp/tests/centrality/betweenness_centrality_reference.hpp b/cpp/tests/centrality/betweenness_centrality_reference.hpp new file mode 100644 index 00000000000..d463f2354cc --- /dev/null +++ b/cpp/tests/centrality/betweenness_centrality_reference.hpp @@ -0,0 +1,232 @@ +/* + * Copyright (c) 2022, NVIDIA CORPORATION. + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governin_from_mtxg permissions and + * limitations under the License. + */ + +#pragma once + +#include + +#include + +#include +#include +#include +#include + +namespace { + +template +void ref_bfs(std::vector const& offsets, + std::vector const& indices, + std::queue& Q, + std::stack& S, + std::vector& dist, + std::vector>& pred, + std::vector& sigmas, + vertex_t source) +{ + pred.clear(); + pred.resize(offsets.size() - 1); + std::fill(dist.begin(), dist.end(), std::numeric_limits::max()); + std::fill(sigmas.begin(), sigmas.end(), double{0}); + dist[source] = 0; + sigmas[source] = 1; + Q.push(source); + + while (!Q.empty()) { + vertex_t v = Q.front(); + Q.pop(); + S.push(v); + for (edge_t nbr_idx = offsets[v]; nbr_idx < offsets[v + 1]; ++nbr_idx) { + vertex_t nbr = indices[nbr_idx]; + // Path Discovery: + // Found for the first time? + if (dist[nbr] == std::numeric_limits::max()) { + dist[nbr] = dist[v] + 1; + Q.push(nbr); + } + // Path counting + // Edge(v, w) on a shortest path? + if (dist[nbr] == dist[v] + 1) { + sigmas[nbr] += sigmas[v]; + pred[nbr].push_back(v); + } + } + } +} + +template +void ref_accumulation(std::vector& result, + std::stack& S, + std::vector>& pred, + std::vector& sigmas, + std::vector& deltas, + vertex_t source) +{ + std::fill(deltas.begin(), deltas.end(), double{0}); + + while (!S.empty()) { + vertex_t w = S.top(); + S.pop(); + for (vertex_t v : pred[w]) { + deltas[v] += (sigmas[v] / sigmas[w]) * (1.0 + deltas[w]); + } + if (w != source) { result[w] += deltas[w]; } + } +} + +template +void ref_endpoints_accumulation(std::vector& result, + std::stack& S, + std::vector>& pred, + std::vector& sigmas, + std::vector& deltas, + vertex_t source) +{ + result[source] += S.size() - 1; + std::fill(deltas.begin(), deltas.end(), double{0}); + + while (!S.empty()) { + vertex_t w = S.top(); + S.pop(); + for (vertex_t v : pred[w]) { + deltas[v] += (sigmas[v] / sigmas[w]) * (1.0 + deltas[w]); + } + if (w != source) { result[w] += deltas[w] + 1; } + } +} + +template +void ref_edge_accumulation(std::vector& result, + std::stack& S, + std::vector>& pred, + std::vector& sigmas, + std::vector& deltas, + vertex_t source) +{ + std::fill(deltas.begin(), deltas.end(), double{0}); + while (!S.empty()) { + vertex_t w = S.top(); + S.pop(); + for (vertex_t v : pred[w]) { + deltas[v] += (sigmas[v] / sigmas[w]) * (1.0 + deltas[w]); + } + if (w != source) { result[w] += deltas[w]; } + } +} + +template +void reference_rescale(result_t* result, + bool directed, + bool normalize, + bool endpoints, + vertex_t const number_of_vertices, + vertex_t const number_of_sources) +{ + bool modified = false; + result_t rescale_factor = static_cast(1); + result_t casted_number_of_sources = static_cast(number_of_sources); + result_t casted_number_of_vertices = static_cast(number_of_vertices); + if (normalize) { + if (number_of_vertices > 2) { + if (endpoints) { + rescale_factor /= (casted_number_of_vertices * (casted_number_of_vertices - 1)); + } else { + rescale_factor /= ((casted_number_of_vertices - 1) * (casted_number_of_vertices - 2)); + } + modified = true; + } + } else { + if (!directed) { + rescale_factor /= static_cast(2); + modified = true; + } + } + if (modified) { + if (number_of_sources > 0) { + rescale_factor *= (casted_number_of_vertices / casted_number_of_sources); + } + } + for (auto idx = 0; idx < number_of_vertices; ++idx) { + result[idx] *= rescale_factor; + } +} + +template +std::vector betweenness_centrality_reference( + std::vector const& offsets, + std::vector const& indices, + std::optional> const& wgt, + std::vector const& seeds, + bool count_endpoints) +{ + std::vector result; + if (offsets.size() > 1) { + result.resize(offsets.size() - 1); + + // Adapted from legacy C++ test implementation + std::queue Q; + std::stack S; + + std::vector dist(result.size()); + std::vector> pred(result.size()); + std::vector sigmas(result.size()); + std::vector deltas(result.size()); + + std::vector neighbors; + + for (vertex_t s : seeds) { + ref_bfs(offsets, indices, Q, S, dist, pred, sigmas, s); + + if (count_endpoints) { + ref_endpoints_accumulation(result, S, pred, sigmas, deltas, s); + } else { + ref_accumulation(result, S, pred, sigmas, deltas, s); + } + } + } + + return result; +} + +template +std::vector edge_betweenness_centrality_reference( + std::vector const& offsets, + std::vector const& indices, + std::optional> const& wgt, + std::vector const& seeds) +{ + std::vector result; + if (indices.size() > 0) { + result.resize(indices.size()); + + // Adapted from legacy C++ test implementation + std::queue Q; + std::stack S; + + std::vector dist(offsets.size() - 1); + std::vector> pred(offsets.size() - 1); + std::vector sigmas(offsets.size() - 1); + std::vector deltas(offsets.size() - 1); + + for (vertex_t s : seeds) { + ref_bfs(offsets, indices, Q, S, dist, pred, sigmas, s); + + ref_edge_accumulation(result, S, pred, sigmas, deltas, s); + } + } + return result; +} +} // namespace diff --git a/cpp/tests/centrality/betweenness_centrality_test.cpp b/cpp/tests/centrality/betweenness_centrality_test.cpp new file mode 100644 index 00000000000..3afc453b7a5 --- /dev/null +++ b/cpp/tests/centrality/betweenness_centrality_test.cpp @@ -0,0 +1,208 @@ +/* + * Copyright (c) 2022, NVIDIA CORPORATION. + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governin_from_mtxg permissions and + * limitations under the License. + */ +#include +#include + +#include +#include +#include +#include +#include + +#include +#include +#include +#include + +#include +#include +#include +#include + +#include + +struct BetweennessCentrality_Usecase { + size_t num_seeds{std::numeric_limits::max()}; + bool normalized{false}; + bool include_endpoints{false}; + bool test_weighted{false}; + bool check_correctness{true}; +}; + +template +class Tests_BetweennessCentrality + : public ::testing::TestWithParam> { + public: + Tests_BetweennessCentrality() {} + + static void SetUpTestCase() {} + static void TearDownTestCase() {} + + virtual void SetUp() {} + virtual void TearDown() {} + + template + void run_current_test(std::tuple const& param) + { + constexpr bool renumber = true; + constexpr bool do_expensive_check = false; + + auto [betweenness_usecase, input_usecase] = param; + + raft::handle_t handle{}; + HighResClock hr_clock{}; + + if (cugraph::test::g_perf) { + RAFT_CUDA_TRY(cudaDeviceSynchronize()); // for consistent performance measurement + hr_clock.start(); + } + + auto [graph, d_renumber_map_labels] = + cugraph::test::construct_graph( + handle, input_usecase, betweenness_usecase.test_weighted, renumber); + + if (cugraph::test::g_perf) { + RAFT_CUDA_TRY(cudaDeviceSynchronize()); // for consistent performance measurement + double elapsed_time{0.0}; + hr_clock.stop(&elapsed_time); + std::cout << "construct_graph took " << elapsed_time * 1e-6 << " s.\n"; + } + + auto graph_view = graph.view(); + + rmm::device_uvector d_seeds(graph_view.number_of_vertices(), handle.get_stream()); + cugraph::detail::sequence_fill( + handle.get_stream(), d_seeds.data(), d_seeds.size(), vertex_t{0}); + + d_seeds = cugraph::test::randomly_select(handle, d_seeds, betweenness_usecase.num_seeds); + + if (cugraph::test::g_perf) { + RAFT_CUDA_TRY(cudaDeviceSynchronize()); // for consistent performance measurement + hr_clock.start(); + } + +#if 0 + auto d_centralities = cugraph::betweenness_centrality( + handle, + graph_view, + std::make_optional>>( + raft::device_span{d_seeds.data(), d_seeds.size()}), + betweenness_usecase.normalized, + betweenness_usecase.include_endpoints, + do_expensive_check); +#else + EXPECT_THROW(cugraph::betweenness_centrality( + handle, + graph_view, + std::make_optional>>( + raft::device_span{d_seeds.data(), d_seeds.size()}), + betweenness_usecase.normalized, + betweenness_usecase.include_endpoints, + do_expensive_check), + cugraph::logic_error); +#endif + + if (cugraph::test::g_perf) { + RAFT_CUDA_TRY(cudaDeviceSynchronize()); // for consistent performance measurement + double elapsed_time{0.0}; + hr_clock.stop(&elapsed_time); + std::cout << "Betweenness Centrality took " << elapsed_time * 1e-6 << " s.\n"; + } + + if (betweenness_usecase.check_correctness) { +#if 0 + auto [h_offsets, h_indices, h_wgt] = cugraph::test::graph_to_host_csr(handle, graph_view); + + auto h_seeds = cugraph::test::to_host(handle, d_seeds); + + auto h_reference_centralities = + betweenness_centrality_reference(h_offsets, h_indices, h_wgt, h_seeds, betweenness_usecase.include_endpoints); + + auto d_reference_centralities = cugraph::test::to_device(handle, h_reference_centralities); + + cugraph::test::betweenness_centrality_validate(handle, + d_renumber_map_labels, + d_centralities, + std::nullopt, + d_reference_centralities); +#endif + } + } +}; + +using Tests_BetweennessCentrality_File = Tests_BetweennessCentrality; +using Tests_BetweennessCentrality_Rmat = Tests_BetweennessCentrality; + +// FIXME: add tests for type combinations +TEST_P(Tests_BetweennessCentrality_File, CheckInt32Int32FloatFloat) +{ + run_current_test( + override_File_Usecase_with_cmd_line_arguments(GetParam())); +} + +TEST_P(Tests_BetweennessCentrality_Rmat, CheckInt32Int32FloatFloat) +{ + run_current_test( + override_Rmat_Usecase_with_cmd_line_arguments(GetParam())); +} + +TEST_P(Tests_BetweennessCentrality_Rmat, CheckInt32Int64FloatFloat) +{ + run_current_test( + override_Rmat_Usecase_with_cmd_line_arguments(GetParam())); +} + +TEST_P(Tests_BetweennessCentrality_Rmat, CheckInt64Int64FloatFloat) +{ + run_current_test( + override_Rmat_Usecase_with_cmd_line_arguments(GetParam())); +} + +INSTANTIATE_TEST_SUITE_P( + file_test_pass, + Tests_BetweennessCentrality_File, + ::testing::Combine( + // enable correctness checks + ::testing::Values(BetweennessCentrality_Usecase{20, false, false, false, true}, + BetweennessCentrality_Usecase{20, false, false, true, true}), + ::testing::Values(cugraph::test::File_Usecase("test/datasets/karate.mtx"), + cugraph::test::File_Usecase("test/datasets/web-Google.mtx"), + cugraph::test::File_Usecase("test/datasets/ljournal-2008.mtx"), + cugraph::test::File_Usecase("test/datasets/webbase-1M.mtx")))); + +INSTANTIATE_TEST_SUITE_P( + rmat_small_test, + Tests_BetweennessCentrality_Rmat, + // enable correctness checks + ::testing::Combine( + ::testing::Values(BetweennessCentrality_Usecase{50, false, false, false, true}, + BetweennessCentrality_Usecase{50, false, false, true, true}), + ::testing::Values(cugraph::test::Rmat_Usecase(10, 16, 0.57, 0.19, 0.19, 0, true, false)))); + +INSTANTIATE_TEST_SUITE_P( + rmat_benchmark_test, /* note that scale & edge factor can be overridden in benchmarking (with + --gtest_filter to select only the rmat_benchmark_test with a specific + vertex & edge type combination) by command line arguments and do not + include more than one Rmat_Usecase that differ only in scale or edge + factor (to avoid running same benchmarks more than once) */ + Tests_BetweennessCentrality_Rmat, + // disable correctness checks for large graphs + ::testing::Combine( + ::testing::Values(BetweennessCentrality_Usecase{500, false, false, false, false}, + BetweennessCentrality_Usecase{500, false, false, true, false}), + ::testing::Values(cugraph::test::Rmat_Usecase(20, 32, 0.57, 0.19, 0.19, 0, false, false)))); + +CUGRAPH_TEST_PROGRAM_MAIN() diff --git a/cpp/tests/centrality/betweenness_centrality_validate.cu b/cpp/tests/centrality/betweenness_centrality_validate.cu new file mode 100644 index 00000000000..697d8714691 --- /dev/null +++ b/cpp/tests/centrality/betweenness_centrality_validate.cu @@ -0,0 +1,131 @@ +/* + * Copyright (c) 2022, NVIDIA CORPORATION. + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governin_from_mtxg permissions and + * limitations under the License. + */ + +#include +#include + +#include + +#include + +namespace cugraph { +namespace test { + +template +void betweenness_centrality_validate( + raft::handle_t const& handle, + std::optional>& d_cugraph_vertex_ids, + rmm::device_uvector& d_cugraph_results, + std::optional>& d_reference_vertex_ids, + rmm::device_uvector& d_reference_results) +{ + auto compare_functor = cugraph::test::nearly_equal{ + weight_t{1e-3}, + weight_t{(weight_t{1} / static_cast(d_cugraph_results.size())) * weight_t{1e-3}}}; + + EXPECT_EQ(d_cugraph_results.size(), d_reference_results.size()); + + if (d_cugraph_vertex_ids) { + thrust::sort_by_key(handle.get_thrust_policy(), + d_cugraph_vertex_ids->begin(), + d_cugraph_vertex_ids->end(), + d_cugraph_results.begin()); + } + + if (d_reference_vertex_ids) { + thrust::sort_by_key(handle.get_thrust_policy(), + d_reference_vertex_ids->begin(), + d_reference_vertex_ids->end(), + d_reference_results.begin()); + } + + EXPECT_TRUE(thrust::equal(handle.get_thrust_policy(), + d_cugraph_results.begin(), + d_cugraph_results.end(), + d_reference_results.begin(), + compare_functor)) + << "Mismatch in centrality results"; +} + +template +void edge_betweenness_centrality_validate(raft::handle_t const& handle, + rmm::device_uvector& d_cugraph_src_vertex_ids, + rmm::device_uvector& d_cugraph_dst_vertex_ids, + rmm::device_uvector& d_cugraph_results, + rmm::device_uvector& d_reference_src_vertex_ids, + rmm::device_uvector& d_reference_dst_vertex_ids, + rmm::device_uvector& d_reference_results) +{ + auto compare_functor = cugraph::test::nearly_equal{ + weight_t{1e-3}, + weight_t{(weight_t{1} / static_cast(d_reference_results.size())) * weight_t{1e-3}}}; + + EXPECT_EQ(d_cugraph_results.size(), d_reference_results.size()); + + thrust::sort_by_key( + handle.get_thrust_policy(), + thrust::make_zip_iterator(d_cugraph_src_vertex_ids.begin(), d_cugraph_dst_vertex_ids.begin()), + thrust::make_zip_iterator(d_cugraph_src_vertex_ids.end(), d_cugraph_dst_vertex_ids.begin()), + d_cugraph_results.begin()); + + thrust::sort_by_key( + handle.get_thrust_policy(), + thrust::make_zip_iterator(d_reference_src_vertex_ids.begin(), + d_reference_dst_vertex_ids.begin()), + thrust::make_zip_iterator(d_reference_src_vertex_ids.end(), d_reference_dst_vertex_ids.begin()), + d_reference_results.begin()); + + EXPECT_TRUE(thrust::equal(handle.get_thrust_policy(), + d_cugraph_results.begin(), + d_cugraph_results.end(), + d_reference_results.begin(), + compare_functor)) + << "Mismatch in centrality results"; +} + +template void betweenness_centrality_validate( + raft::handle_t const& handle, + std::optional>& d_cugraph_vertex_ids, + rmm::device_uvector& d_cugraph_results, + std::optional>& d_reference_vertex_ids, + rmm::device_uvector& d_reference_results); + +template void betweenness_centrality_validate( + raft::handle_t const& handle, + std::optional>& d_cugraph_vertex_ids, + rmm::device_uvector& d_cugraph_results, + std::optional>& d_reference_vertex_ids, + rmm::device_uvector& d_reference_results); + +template void edge_betweenness_centrality_validate( + raft::handle_t const& handle, + rmm::device_uvector& d_cugraph_src_vertex_ids, + rmm::device_uvector& d_cugraph_dst_vertex_ids, + rmm::device_uvector& d_cugraph_results, + rmm::device_uvector& d_reference_src_vertex_ids, + rmm::device_uvector& d_reference_dst_vertex_ids, + rmm::device_uvector& d_reference_results); +template void edge_betweenness_centrality_validate( + raft::handle_t const& handle, + rmm::device_uvector& d_cugraph_src_vertex_ids, + rmm::device_uvector& d_cugraph_dst_vertex_ids, + rmm::device_uvector& d_cugraph_results, + rmm::device_uvector& d_reference_src_vertex_ids, + rmm::device_uvector& d_reference_dst_vertex_ids, + rmm::device_uvector& d_reference_results); + +} // namespace test +} // namespace cugraph diff --git a/cpp/tests/centrality/betweenness_centrality_validate.hpp b/cpp/tests/centrality/betweenness_centrality_validate.hpp new file mode 100644 index 00000000000..48d53594ba9 --- /dev/null +++ b/cpp/tests/centrality/betweenness_centrality_validate.hpp @@ -0,0 +1,45 @@ +/* + * Copyright (c) 2022, NVIDIA CORPORATION. + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governin_from_mtxg permissions and + * limitations under the License. + */ +#pragma once + +#include + +#include + +#include + +namespace cugraph { +namespace test { + +template +void betweenness_centrality_validate( + raft::handle_t const& handle, + std::optional>& d_cugraph_vertex_ids, + rmm::device_uvector& d_cugraph_results, + std::optional>& d_reference_vertex_ids, + rmm::device_uvector& d_reference_results); + +template +void edge_betweenness_centrality_validate(raft::handle_t const& handle, + rmm::device_uvector& d_cugraph_src_vertex_ids, + rmm::device_uvector& d_cugraph_dst_vertex_ids, + rmm::device_uvector& d_cugraph_results, + rmm::device_uvector& d_reference_src_vertex_ids, + rmm::device_uvector& d_reference_dst_vertex_ids, + rmm::device_uvector& d_reference_results); + +} // namespace test +} // namespace cugraph diff --git a/cpp/tests/centrality/edge_betweenness_centrality_test.cpp b/cpp/tests/centrality/edge_betweenness_centrality_test.cpp new file mode 100644 index 00000000000..68d7cbb9f86 --- /dev/null +++ b/cpp/tests/centrality/edge_betweenness_centrality_test.cpp @@ -0,0 +1,210 @@ +/* + * Copyright (c) 2022, NVIDIA CORPORATION. + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governin_from_mtxg permissions and + * limitations under the License. + */ +#include +#include + +#include +#include +#include +#include +#include + +#include +#include +#include +#include + +#include +#include +#include +#include + +#include + +struct EdgeBetweennessCentrality_Usecase { + size_t num_seeds{std::numeric_limits::max()}; + bool normalized{false}; + bool test_weighted{false}; + bool check_correctness{true}; +}; + +template +class Tests_EdgeBetweennessCentrality + : public ::testing::TestWithParam< + std::tuple> { + public: + Tests_EdgeBetweennessCentrality() {} + + static void SetUpTestCase() {} + static void TearDownTestCase() {} + + virtual void SetUp() {} + virtual void TearDown() {} + + template + void run_current_test(std::tuple const& param) + { + constexpr bool renumber = true; + constexpr bool do_expensive_check = false; + + auto [betweenness_usecase, input_usecase] = param; + + raft::handle_t handle{}; + HighResClock hr_clock{}; + + if (cugraph::test::g_perf) { + RAFT_CUDA_TRY(cudaDeviceSynchronize()); // for consistent performance measurement + hr_clock.start(); + } + + auto [graph, d_renumber_map_labels] = + cugraph::test::construct_graph( + handle, input_usecase, betweenness_usecase.test_weighted, renumber); + + if (cugraph::test::g_perf) { + RAFT_CUDA_TRY(cudaDeviceSynchronize()); // for consistent performance measurement + double elapsed_time{0.0}; + hr_clock.stop(&elapsed_time); + std::cout << "construct_graph took " << elapsed_time * 1e-6 << " s.\n"; + } + + auto graph_view = graph.view(); + + rmm::device_uvector d_seeds(graph_view.number_of_vertices(), handle.get_stream()); + cugraph::detail::sequence_fill( + handle.get_stream(), d_seeds.data(), d_seeds.size(), vertex_t{0}); + + d_seeds = cugraph::test::randomly_select(handle, d_seeds, betweenness_usecase.num_seeds); + + if (cugraph::test::g_perf) { + RAFT_CUDA_TRY(cudaDeviceSynchronize()); // for consistent performance measurement + hr_clock.start(); + } + +#if 0 + auto d_centralities = cugraph::edge_betweenness_centrality( + handle, + graph_view, + std::make_optional>>( + raft::device_span{d_seeds.data(), d_seeds.size()}), + betweenness_usecase.normalized, + do_expensive_check); +#else + EXPECT_THROW(cugraph::edge_betweenness_centrality( + handle, + graph_view, + std::make_optional>>( + raft::device_span{d_seeds.data(), d_seeds.size()}), + betweenness_usecase.normalized, + do_expensive_check), + cugraph::logic_error); +#endif + + if (cugraph::test::g_perf) { + RAFT_CUDA_TRY(cudaDeviceSynchronize()); // for consistent performance measurement + double elapsed_time{0.0}; + hr_clock.stop(&elapsed_time); + std::cout << "Edge Betweenness Centrality took " << elapsed_time * 1e-6 << " s.\n"; + } + + if (betweenness_usecase.check_correctness) { +#if 0 + auto [h_offsets, h_indices, h_wgt] = cugraph::test::graph_to_host_csr(handle, graph_view); + + auto h_seeds = cugraph::test::to_host(handle, d_seeds); + + auto h_reference_centralities = + betweenness_centrality_reference(h_offsets, h_indices, h_wgt, h_seeds, betweenness_usecase.include_endpoints); + + auto d_reference_centralities = cugraph::test::to_device(handle, h_reference_centralities); + + // Need to get edges in order... + + cugraph::test::edge_betweenness_centrality_validate(handle, + d_renumber_map_labels, + d_centralities, + std::nullopt, + d_reference_centralities); +#endif + } + } +}; + +using Tests_EdgeBetweennessCentrality_File = + Tests_EdgeBetweennessCentrality; +using Tests_EdgeBetweennessCentrality_Rmat = + Tests_EdgeBetweennessCentrality; + +// FIXME: add tests for type combinations +TEST_P(Tests_EdgeBetweennessCentrality_File, CheckInt32Int32FloatFloat) +{ + run_current_test( + override_File_Usecase_with_cmd_line_arguments(GetParam())); +} + +TEST_P(Tests_EdgeBetweennessCentrality_Rmat, CheckInt32Int32FloatFloat) +{ + run_current_test( + override_Rmat_Usecase_with_cmd_line_arguments(GetParam())); +} + +TEST_P(Tests_EdgeBetweennessCentrality_Rmat, CheckInt32Int64FloatFloat) +{ + run_current_test( + override_Rmat_Usecase_with_cmd_line_arguments(GetParam())); +} + +TEST_P(Tests_EdgeBetweennessCentrality_Rmat, CheckInt64Int64FloatFloat) +{ + run_current_test( + override_Rmat_Usecase_with_cmd_line_arguments(GetParam())); +} + +INSTANTIATE_TEST_SUITE_P( + file_test_pass, + Tests_EdgeBetweennessCentrality_File, + ::testing::Combine( + // enable correctness checks + ::testing::Values(EdgeBetweennessCentrality_Usecase{20, false, false, true}, + EdgeBetweennessCentrality_Usecase{20, false, true, true}), + ::testing::Values(cugraph::test::File_Usecase("test/datasets/karate.mtx"), + cugraph::test::File_Usecase("test/datasets/web-Google.mtx"), + cugraph::test::File_Usecase("test/datasets/ljournal-2008.mtx"), + cugraph::test::File_Usecase("test/datasets/webbase-1M.mtx")))); + +INSTANTIATE_TEST_SUITE_P( + rmat_small_test, + Tests_EdgeBetweennessCentrality_Rmat, + // enable correctness checks + ::testing::Combine( + ::testing::Values(EdgeBetweennessCentrality_Usecase{50, false, false, true}, + EdgeBetweennessCentrality_Usecase{50, false, true, true}), + ::testing::Values(cugraph::test::Rmat_Usecase(10, 16, 0.57, 0.19, 0.19, 0, true, false)))); + +INSTANTIATE_TEST_SUITE_P( + rmat_benchmark_test, /* note that scale & edge factor can be overridden in benchmarking (with + --gtest_filter to select only the rmat_benchmark_test with a specific + vertex & edge type combination) by command line arguments and do not + include more than one Rmat_Usecase that differ only in scale or edge + factor (to avoid running same benchmarks more than once) */ + Tests_EdgeBetweennessCentrality_Rmat, + // disable correctness checks for large graphs + ::testing::Combine( + ::testing::Values(EdgeBetweennessCentrality_Usecase{500, false, false, false}, + EdgeBetweennessCentrality_Usecase{500, false, true, false}), + ::testing::Values(cugraph::test::Rmat_Usecase(20, 32, 0.57, 0.19, 0.19, 0, false, false)))); + +CUGRAPH_TEST_PROGRAM_MAIN() diff --git a/cpp/tests/centrality/mg_betweenness_centrality_test.cpp b/cpp/tests/centrality/mg_betweenness_centrality_test.cpp new file mode 100644 index 00000000000..6679b1c67c8 --- /dev/null +++ b/cpp/tests/centrality/mg_betweenness_centrality_test.cpp @@ -0,0 +1,232 @@ +/* + * Copyright (c) 2022, NVIDIA CORPORATION. + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governin_from_mtxg permissions and + * limitations under the License. + */ +#include + +#include +#include +#include +#include +#include +#include + +#include +#include +#include +#include + +#include +#include +#include +#include + +#include + +struct BetweennessCentrality_Usecase { + size_t num_seeds{std::numeric_limits::max()}; + bool normalized{false}; + bool include_endpoints{false}; + bool test_weighted{false}; + bool check_correctness{true}; +}; + +template +class Tests_MGBetweennessCentrality + : public ::testing::TestWithParam> { + public: + Tests_MGBetweennessCentrality() {} + + static void SetUpTestCase() { handle_ = cugraph::test::initialize_mg_handle(); } + static void TearDownTestCase() { handle_.reset(); } + + virtual void SetUp() {} + virtual void TearDown() {} + + template + void run_current_test(std::tuple const& param) + { + constexpr bool renumber = true; + constexpr bool do_expensive_check = false; + + auto [betweenness_usecase, input_usecase] = param; + + HighResClock hr_clock{}; + + if (cugraph::test::g_perf) { + RAFT_CUDA_TRY(cudaDeviceSynchronize()); // for consistent performance measurement + hr_clock.start(); + } + + auto [mg_graph, mg_renumber_map] = + cugraph::test::construct_graph( + *handle_, input_usecase, betweenness_usecase.test_weighted, renumber); + + if (cugraph::test::g_perf) { + RAFT_CUDA_TRY(cudaDeviceSynchronize()); // for consistent performance measurement + double elapsed_time{0.0}; + hr_clock.stop(&elapsed_time); + std::cout << "construct_graph took " << elapsed_time * 1e-6 << " s.\n"; + } + + auto mg_graph_view = mg_graph.view(); + + rmm::device_uvector d_seeds(0, handle_->get_stream()); + + if (handle_->get_comms().get_rank() == 0) { + rmm::device_uvector d_seeds(mg_graph_view.number_of_vertices(), + handle_->get_stream()); + cugraph::detail::sequence_fill( + handle_->get_stream(), d_seeds.data(), d_seeds.size(), vertex_t{0}); + + d_seeds = cugraph::test::randomly_select(*handle_, d_seeds, betweenness_usecase.num_seeds); + } + + d_seeds = cugraph::detail::shuffle_ext_vertices_by_gpu_id(*handle_, std::move(d_seeds)); + + if (cugraph::test::g_perf) { + RAFT_CUDA_TRY(cudaDeviceSynchronize()); // for consistent performance measurement + hr_clock.start(); + } + +#if 0 + auto d_centralities = cugraph::betweenness_centrality( + *handle_, + mg_graph_view, + std::make_optional>>( + raft::device_span{d_seeds.data(), d_seeds.size()}), + betweenness_usecase.normalized, + betweenness_usecase.include_endpoints, + do_expensive_check); +#else + EXPECT_THROW(cugraph::betweenness_centrality( + *handle_, + mg_graph_view, + std::make_optional>>( + raft::device_span{d_seeds.data(), d_seeds.size()}), + betweenness_usecase.normalized, + betweenness_usecase.include_endpoints, + do_expensive_check), + cugraph::logic_error); +#endif + + if (cugraph::test::g_perf) { + RAFT_CUDA_TRY(cudaDeviceSynchronize()); // for consistent performance measurement + double elapsed_time{0.0}; + hr_clock.stop(&elapsed_time); + std::cout << "Betweenness Centrality took " << elapsed_time * 1e-6 << " s.\n"; + } + + if (betweenness_usecase.check_correctness) { +#if 0 + d_centralities = cugraph::test::device_gatherv( + *handle_, raft::device_span{d_centralities.data(), d_centralities.size()}); + d_seeds = cugraph::test::device_gatherv( + *handle_, raft::device_span{d_seeds.data(), d_seeds.size()}); + + auto [sg_graph, sg_renumber_map] = + cugraph::test::mg_graph_to_sg_graph(*handle_, mg_graph_view, mg_renumber_map, true); + + auto d_reference_centralities = cugraph::betweenness_centrality( + *handle_, + sg_graph.view(), + std::optional{std::nullopt}, + std::make_optional>(d_seeds.data(), d_seeds.size()), + betweenness_usecase.normalized, + betweenness_usecase.include_endpoints, + do_expensive_check); + + if (d_seeds.size() > 0) { + cugraph::test::betweenness_centrality_validate(handle, + mg_renumber_map, + d_centralities, + sg_renumber_map, + d_reference_centralities); + } +#endif + } + } + + private: + static std::unique_ptr handle_; +}; + +template +std::unique_ptr Tests_MGBetweennessCentrality::handle_ = nullptr; + +using Tests_MGBetweennessCentrality_File = + Tests_MGBetweennessCentrality; +using Tests_MGBetweennessCentrality_Rmat = + Tests_MGBetweennessCentrality; + +// FIXME: add tests for type combinations +TEST_P(Tests_MGBetweennessCentrality_File, CheckInt32Int32FloatFloat) +{ + run_current_test( + override_File_Usecase_with_cmd_line_arguments(GetParam())); +} + +TEST_P(Tests_MGBetweennessCentrality_Rmat, CheckInt32Int32FloatFloat) +{ + run_current_test( + override_Rmat_Usecase_with_cmd_line_arguments(GetParam())); +} + +TEST_P(Tests_MGBetweennessCentrality_Rmat, CheckInt32Int64FloatFloat) +{ + run_current_test( + override_Rmat_Usecase_with_cmd_line_arguments(GetParam())); +} + +TEST_P(Tests_MGBetweennessCentrality_Rmat, CheckInt64Int64FloatFloat) +{ + run_current_test( + override_Rmat_Usecase_with_cmd_line_arguments(GetParam())); +} + +INSTANTIATE_TEST_SUITE_P( + file_test_pass, + Tests_MGBetweennessCentrality_File, + ::testing::Combine( + // enable correctness checks + ::testing::Values(BetweennessCentrality_Usecase{20, false, false, false, true}, + BetweennessCentrality_Usecase{20, false, false, true, true}), + ::testing::Values(cugraph::test::File_Usecase("test/datasets/karate.mtx"), + cugraph::test::File_Usecase("test/datasets/web-Google.mtx"), + cugraph::test::File_Usecase("test/datasets/ljournal-2008.mtx"), + cugraph::test::File_Usecase("test/datasets/webbase-1M.mtx")))); + +INSTANTIATE_TEST_SUITE_P( + rmat_small_test, + Tests_MGBetweennessCentrality_Rmat, + // enable correctness checks + ::testing::Combine( + ::testing::Values(BetweennessCentrality_Usecase{50, false, false, false, true}, + BetweennessCentrality_Usecase{50, false, false, true, true}), + ::testing::Values(cugraph::test::Rmat_Usecase(10, 16, 0.57, 0.19, 0.19, 0, true, false)))); + +INSTANTIATE_TEST_SUITE_P( + rmat_benchmark_test, /* note that scale & edge factor can be overridden in benchmarking (with + --gtest_filter to select only the rmat_benchmark_test with a specific + vertex & edge type combination) by command line arguments and do not + include more than one Rmat_Usecase that differ only in scale or edge + factor (to avoid running same benchmarks more than once) */ + Tests_MGBetweennessCentrality_Rmat, + // disable correctness checks for large graphs + ::testing::Combine( + ::testing::Values(BetweennessCentrality_Usecase{500, false, false, false, false}, + BetweennessCentrality_Usecase{500, false, false, true, false}), + ::testing::Values(cugraph::test::Rmat_Usecase(20, 32, 0.57, 0.19, 0.19, 0, false, false)))); + +CUGRAPH_MG_TEST_PROGRAM_MAIN() diff --git a/cpp/tests/centrality/mg_edge_betweenness_centrality_test.cpp b/cpp/tests/centrality/mg_edge_betweenness_centrality_test.cpp new file mode 100644 index 00000000000..9dfb6606f7c --- /dev/null +++ b/cpp/tests/centrality/mg_edge_betweenness_centrality_test.cpp @@ -0,0 +1,220 @@ +/* + * Copyright (c) 2022, NVIDIA CORPORATION. + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governin_from_mtxg permissions and + * limitations under the License. + */ +#include + +#include +#include +#include +#include +#include +#include + +#include +#include +#include +#include + +#include +#include +#include +#include + +#include + +struct EdgeBetweennessCentrality_Usecase { + size_t num_seeds{std::numeric_limits::max()}; + bool normalized{false}; + bool test_weighted{false}; + bool check_correctness{true}; +}; + +template +class Tests_MGEdgeBetweennessCentrality + : public ::testing::TestWithParam< + std::tuple> { + public: + Tests_MGEdgeBetweennessCentrality() {} + + static void SetUpTestCase() { handle_ = cugraph::test::initialize_mg_handle(); } + static void TearDownTestCase() { handle_.reset(); } + + virtual void SetUp() {} + virtual void TearDown() {} + + template + void run_current_test(std::tuple const& param) + { + constexpr bool renumber = true; + constexpr bool do_expensive_check = false; + + auto [betweenness_usecase, input_usecase] = param; + + HighResClock hr_clock{}; + + if (cugraph::test::g_perf) { + RAFT_CUDA_TRY(cudaDeviceSynchronize()); // for consistent performance measurement + hr_clock.start(); + } + + auto [graph, d_renumber_map_labels] = + cugraph::test::construct_graph( + *handle_, input_usecase, betweenness_usecase.test_weighted, renumber); + + if (cugraph::test::g_perf) { + RAFT_CUDA_TRY(cudaDeviceSynchronize()); // for consistent performance measurement + double elapsed_time{0.0}; + hr_clock.stop(&elapsed_time); + std::cout << "construct_graph took " << elapsed_time * 1e-6 << " s.\n"; + } + + auto graph_view = graph.view(); + + rmm::device_uvector d_seeds(0, handle_->get_stream()); + + if (handle_->get_comms().get_rank() == 0) { + rmm::device_uvector d_seeds(graph_view.number_of_vertices(), handle_->get_stream()); + cugraph::detail::sequence_fill( + handle_->get_stream(), d_seeds.data(), d_seeds.size(), vertex_t{0}); + + d_seeds = cugraph::test::randomly_select(*handle_, d_seeds, betweenness_usecase.num_seeds); + } + + d_seeds = cugraph::detail::shuffle_ext_vertices_by_gpu_id(*handle_, std::move(d_seeds)); + + if (cugraph::test::g_perf) { + RAFT_CUDA_TRY(cudaDeviceSynchronize()); // for consistent performance measurement + hr_clock.start(); + } + +#if 0 + auto d_centralities = cugraph::edge_betweenness_centrality( + *handle_, + graph_view, + std::make_optional>>( + raft::device_span{d_seeds.data(), d_seeds.size()}), + betweenness_usecase.normalized, + do_expensive_check); +#else + EXPECT_THROW(cugraph::edge_betweenness_centrality( + *handle_, + graph_view, + std::make_optional>>( + raft::device_span{d_seeds.data(), d_seeds.size()}), + betweenness_usecase.normalized, + do_expensive_check), + cugraph::logic_error); +#endif + + if (cugraph::test::g_perf) { + RAFT_CUDA_TRY(cudaDeviceSynchronize()); // for consistent performance measurement + double elapsed_time{0.0}; + hr_clock.stop(&elapsed_time); + std::cout << "Edge Betweenness Centrality took " << elapsed_time * 1e-6 << " s.\n"; + } + + if (betweenness_usecase.check_correctness) { +#if 0 + d_centralities = cugraph::test::device_gatherv( + *handle_, raft::device_span(d_centralities.data(), d_centralities.size())); + d_seeds = cugraph::test::device_gatherv( + *handle_, raft::device_span(d_seeds.data(), d_seeds.size())); + + auto [h_src, h_dst, h_wgt] = cugraph::test::graph_to_host_coo(*handle_, graph_view); + + if (h_src.size() > 0) { + auto h_centralities = cugraph::test::to_host(*handle_, d_centralities); + auto h_seeds = cugraph::test::to_host(*handle_, d_seeds); + + cugraph::test::edge_betweenness_centrality_validate( + h_src, h_dst, h_wgt, h_centralities, h_seeds); + } +#endif + } + } + + private: + static std::unique_ptr handle_; +}; + +template +std::unique_ptr Tests_MGEdgeBetweennessCentrality::handle_ = + nullptr; + +using Tests_MGEdgeBetweennessCentrality_File = + Tests_MGEdgeBetweennessCentrality; +using Tests_MGEdgeBetweennessCentrality_Rmat = + Tests_MGEdgeBetweennessCentrality; + +// FIXME: add tests for type combinations +TEST_P(Tests_MGEdgeBetweennessCentrality_File, CheckInt32Int32FloatFloat) +{ + run_current_test( + override_File_Usecase_with_cmd_line_arguments(GetParam())); +} + +TEST_P(Tests_MGEdgeBetweennessCentrality_Rmat, CheckInt32Int32FloatFloat) +{ + run_current_test( + override_Rmat_Usecase_with_cmd_line_arguments(GetParam())); +} + +TEST_P(Tests_MGEdgeBetweennessCentrality_Rmat, CheckInt32Int64FloatFloat) +{ + run_current_test( + override_Rmat_Usecase_with_cmd_line_arguments(GetParam())); +} + +TEST_P(Tests_MGEdgeBetweennessCentrality_Rmat, CheckInt64Int64FloatFloat) +{ + run_current_test( + override_Rmat_Usecase_with_cmd_line_arguments(GetParam())); +} + +INSTANTIATE_TEST_SUITE_P( + file_test_pass, + Tests_MGEdgeBetweennessCentrality_File, + ::testing::Combine( + // enable correctness checks + ::testing::Values(EdgeBetweennessCentrality_Usecase{20, false, false, true}, + EdgeBetweennessCentrality_Usecase{20, false, true, true}), + ::testing::Values(cugraph::test::File_Usecase("test/datasets/karate.mtx"), + cugraph::test::File_Usecase("test/datasets/web-Google.mtx"), + cugraph::test::File_Usecase("test/datasets/ljournal-2008.mtx"), + cugraph::test::File_Usecase("test/datasets/webbase-1M.mtx")))); + +INSTANTIATE_TEST_SUITE_P( + rmat_small_test, + Tests_MGEdgeBetweennessCentrality_Rmat, + // enable correctness checks + ::testing::Combine( + ::testing::Values(EdgeBetweennessCentrality_Usecase{50, false, false, true}, + EdgeBetweennessCentrality_Usecase{50, false, true, true}), + ::testing::Values(cugraph::test::Rmat_Usecase(10, 16, 0.57, 0.19, 0.19, 0, true, false)))); + +INSTANTIATE_TEST_SUITE_P( + rmat_benchmark_test, /* note that scale & edge factor can be overridden in benchmarking (with + --gtest_filter to select only the rmat_benchmark_test with a specific + vertex & edge type combination) by command line arguments and do not + include more than one Rmat_Usecase that differ only in scale or edge + factor (to avoid running same benchmarks more than once) */ + Tests_MGEdgeBetweennessCentrality_Rmat, + // disable correctness checks for large graphs + ::testing::Combine( + ::testing::Values(EdgeBetweennessCentrality_Usecase{500, false, false, false}, + EdgeBetweennessCentrality_Usecase{500, false, true, false}), + ::testing::Values(cugraph::test::Rmat_Usecase(20, 32, 0.57, 0.19, 0.19, 0, false, false)))); + +CUGRAPH_MG_TEST_PROGRAM_MAIN()