From fdc6aa5acde918e50b6f6f365cc1ec03fcfb18ec Mon Sep 17 00:00:00 2001 From: Chuck Hastings <45364586+ChuckHastings@users.noreply.github.com> Date: Tue, 12 Mar 2024 19:01:26 -0400 Subject: [PATCH] Add degrees to C API (#4212) Add new method `cugraph_degrees` to the C API to compute and return the degrees of vertices. Closes #4171 Authors: - Chuck Hastings (https://github.com/ChuckHastings) - Joseph Nke (https://github.com/jnke2016) - Ralph Liu (https://github.com/nv-rliu) Approvers: - Seunghwa Kang (https://github.com/seunghwak) - Rick Ratzel (https://github.com/rlratzel) URL: https://github.com/rapidsai/cugraph/pull/4212 --- cpp/CMakeLists.txt | 2 + cpp/include/cugraph_c/graph_functions.h | 112 +++++ cpp/src/c_api/abstract_functor.hpp | 2 +- cpp/src/c_api/degrees.cu | 225 ++++++++++ cpp/src/c_api/degrees_result.cpp | 63 +++ cpp/src/c_api/degrees_result.hpp | 32 ++ cpp/tests/CMakeLists.txt | 2 + cpp/tests/c_api/c_test_utils.h | 4 +- cpp/tests/c_api/degrees_test.c | 387 +++++++++++++++++ cpp/tests/c_api/mg_degrees_test.c | 407 ++++++++++++++++++ cpp/tests/c_api/test_utils.cpp | 14 +- .../simpleDistributedGraph.py | 307 +++++++------ .../graph_implementation/simpleGraph.py | 198 +++++---- .../centrality/test_degree_centrality_mg.py | 4 +- .../pylibcugraph/pylibcugraph/CMakeLists.txt | 3 +- python/pylibcugraph/pylibcugraph/__init__.py | 2 + .../_cugraph_c/graph_functions.pxd | 55 +++ python/pylibcugraph/pylibcugraph/degrees.pyx | 307 +++++++++++++ 18 files changed, 1905 insertions(+), 221 deletions(-) create mode 100644 cpp/src/c_api/degrees.cu create mode 100644 cpp/src/c_api/degrees_result.cpp create mode 100644 cpp/src/c_api/degrees_result.hpp create mode 100644 cpp/tests/c_api/degrees_test.c create mode 100644 cpp/tests/c_api/mg_degrees_test.c create mode 100644 python/pylibcugraph/pylibcugraph/degrees.pyx diff --git a/cpp/CMakeLists.txt b/cpp/CMakeLists.txt index b12403710ab..88908ef70ce 100644 --- a/cpp/CMakeLists.txt +++ b/cpp/CMakeLists.txt @@ -415,6 +415,8 @@ endif() add_library(cugraph_c src/c_api/resource_handle.cpp src/c_api/array.cpp + src/c_api/degrees.cu + src/c_api/degrees_result.cpp src/c_api/error.cpp src/c_api/graph_sg.cpp src/c_api/graph_mg.cpp diff --git a/cpp/include/cugraph_c/graph_functions.h b/cpp/include/cugraph_c/graph_functions.h index 8fe1ea0b958..94b06189796 100644 --- a/cpp/include/cugraph_c/graph_functions.h +++ b/cpp/include/cugraph_c/graph_functions.h @@ -229,6 +229,118 @@ cugraph_error_code_t cugraph_allgather(const cugraph_resource_handle_t* handle, cugraph_induced_subgraph_result_t** result, cugraph_error_t** error); +/** + * @brief Opaque degree result type + */ +typedef struct { + int32_t align_; +} cugraph_degrees_result_t; + +/** + * @brief Compute in degrees + * + * Compute the in degrees for the vertices in the graph. + * + * @param [in] handle Handle for accessing resources. + * @param [in] graph Pointer to graph + * @param [in] source_vertices Device array of vertices we want to compute in degrees for. + * @param [in] do_expensive_check A flag to run expensive checks for input arguments (if set to + * true) + * @param [out] result Opaque pointer to degrees result + * @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_in_degrees( + const cugraph_resource_handle_t* handle, + cugraph_graph_t* graph, + const cugraph_type_erased_device_array_view_t* source_vertices, + bool_t do_expensive_check, + cugraph_degrees_result_t** result, + cugraph_error_t** error); + +/** + * @brief Compute out degrees + * + * Compute the out degrees for the vertices in the graph. + * + * @param [in] handle Handle for accessing resources. + * @param [in] graph Pointer to graph + * @param [in] source_vertices Device array of vertices we want to compute out degrees for. + * @param [in] do_expensive_check A flag to run expensive checks for input arguments (if set to + * true) + * @param [out] result Opaque pointer to degrees result + * @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_out_degrees( + const cugraph_resource_handle_t* handle, + cugraph_graph_t* graph, + const cugraph_type_erased_device_array_view_t* source_vertices, + bool_t do_expensive_check, + cugraph_degrees_result_t** result, + cugraph_error_t** error); + +/** + * @brief Compute degrees + * + * Compute the degrees for the vertices in the graph. + * + * @param [in] handle Handle for accessing resources. + * @param [in] graph Pointer to graph + * @param [in] source_vertices Device array of vertices we want to compute degrees for. + * @param [in] do_expensive_check A flag to run expensive checks for input arguments (if set to + * true) + * @param [out] result Opaque pointer to degrees result + * @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_degrees(const cugraph_resource_handle_t* handle, + cugraph_graph_t* graph, + const cugraph_type_erased_device_array_view_t* source_vertices, + bool_t do_expensive_check, + cugraph_degrees_result_t** result, + cugraph_error_t** error); + +/** + * @brief Get the vertex ids + * + * @param [in] degrees_result Opaque pointer to degree result + * @return type erased array view of vertex ids + */ +cugraph_type_erased_device_array_view_t* cugraph_degrees_result_get_vertices( + cugraph_degrees_result_t* degrees_result); + +/** + * @brief Get the in degrees + * + * @param [in] degrees_result Opaque pointer to degree result + * @return type erased array view of vertex ids + */ +cugraph_type_erased_device_array_view_t* cugraph_degrees_result_get_in_degrees( + cugraph_degrees_result_t* degrees_result); + +/** + * @brief Get the out degrees + * + * If the graph is symmetric, in degrees and out degrees will be equal (and + * will be stored in the same memory). + * + * @param [in] degrees_result Opaque pointer to degree result + * @return type erased array view of vertex ids + */ +cugraph_type_erased_device_array_view_t* cugraph_degrees_result_get_out_degrees( + cugraph_degrees_result_t* degrees_result); + +/** + * @brief Free degree result + * + * @param [in] degrees_result Opaque pointer to degree result + */ +void cugraph_degrees_result_free(cugraph_degrees_result_t* degrees_result); + #ifdef __cplusplus } #endif diff --git a/cpp/src/c_api/abstract_functor.hpp b/cpp/src/c_api/abstract_functor.hpp index 219b1256065..8d3ed11341f 100644 --- a/cpp/src/c_api/abstract_functor.hpp +++ b/cpp/src/c_api/abstract_functor.hpp @@ -27,7 +27,7 @@ namespace c_api { struct abstract_functor { // Move to abstract functor... make operator a void, add cugraph_graph_t * result to functor // try that with instantiation questions - std::unique_ptr error_{std::make_unique("")}; + std::unique_ptr error_ = {std::make_unique("")}; cugraph_error_code_t error_code_{CUGRAPH_SUCCESS}; void unsupported() diff --git a/cpp/src/c_api/degrees.cu b/cpp/src/c_api/degrees.cu new file mode 100644 index 00000000000..d6481efa905 --- /dev/null +++ b/cpp/src/c_api/degrees.cu @@ -0,0 +1,225 @@ +/* + * Copyright (c) 2024, 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_api/abstract_functor.hpp" +#include "c_api/degrees_result.hpp" +#include "c_api/graph.hpp" +#include "c_api/resource_handle.hpp" +#include "c_api/utils.hpp" + +#include + +#include +#include +#include +#include +#include + +#include + +#include + +namespace { + +struct degrees_functor : public cugraph::c_api::abstract_functor { + raft::handle_t const& handle_; + cugraph::c_api::cugraph_graph_t* graph_{}; + cugraph::c_api::cugraph_type_erased_device_array_view_t const* source_vertices_; + bool in_degrees_{false}; + bool out_degrees_{false}; + bool do_expensive_check_{false}; + cugraph::c_api::cugraph_degrees_result_t* result_{}; + + degrees_functor(cugraph_resource_handle_t const* handle, + cugraph_graph_t* graph, + ::cugraph_type_erased_device_array_view_t const* source_vertices, + bool in_degrees, + bool out_degrees, + bool do_expensive_check) + : abstract_functor(), + handle_(*reinterpret_cast(handle)->handle_), + graph_(reinterpret_cast(graph)), + source_vertices_( + reinterpret_cast( + source_vertices)), + in_degrees_{in_degrees}, + out_degrees_{out_degrees}, + 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 { + auto graph = + reinterpret_cast*>( + graph_->graph_); + + auto graph_view = graph->view(); + + auto number_map = reinterpret_cast*>(graph_->number_map_); + + std::optional> in_degrees{std::nullopt}; + std::optional> out_degrees{std::nullopt}; + + if (in_degrees_ && out_degrees_ && graph_view.is_symmetric()) { + in_degrees = store_transposed ? graph_view.compute_in_degrees(handle_) + : graph_view.compute_out_degrees(handle_); + // out_degrees will be extracted from in_degrees in the result + } else { + if (in_degrees_) in_degrees = graph_view.compute_in_degrees(handle_); + + if (out_degrees_) out_degrees = graph_view.compute_out_degrees(handle_); + } + + rmm::device_uvector vertex_ids(0, handle_.get_stream()); + + if (source_vertices_) { + // FIXME: Would be more efficient if graph_view.compute_*_degrees could take a vertex + // subset + vertex_ids.resize(source_vertices_->size_, handle_.get_stream()); + raft::copy(vertex_ids.data(), + source_vertices_->as_type(), + vertex_ids.size(), + handle_.get_stream()); + + if constexpr (multi_gpu) { + vertex_ids = cugraph::detail::shuffle_ext_vertices_to_local_gpu_by_vertex_partitioning( + handle_, std::move(vertex_ids)); + } + + cugraph::renumber_ext_vertices( + handle_, + vertex_ids.data(), + vertex_ids.size(), + number_map->data(), + graph_view.local_vertex_partition_range_first(), + graph_view.local_vertex_partition_range_last(), + do_expensive_check_); + + auto vertex_partition = cugraph::vertex_partition_device_view_t( + graph_view.local_vertex_partition_view()); + + auto vertices_iter = thrust::make_transform_iterator( + vertex_ids.begin(), + cuda::proclaim_return_type([vertex_partition] __device__(auto v) { + return vertex_partition.local_vertex_partition_offset_from_vertex_nocheck(v); + })); + + if (in_degrees && out_degrees) { + rmm::device_uvector tmp_in_degrees(vertex_ids.size(), handle_.get_stream()); + rmm::device_uvector tmp_out_degrees(vertex_ids.size(), handle_.get_stream()); + thrust::gather( + handle_.get_thrust_policy(), + vertices_iter, + vertices_iter + vertex_ids.size(), + thrust::make_zip_iterator(in_degrees->begin(), out_degrees->begin()), + thrust::make_zip_iterator(tmp_in_degrees.begin(), tmp_out_degrees.begin())); + in_degrees = std::move(tmp_in_degrees); + out_degrees = std::move(tmp_out_degrees); + } else if (in_degrees) { + rmm::device_uvector tmp_in_degrees(vertex_ids.size(), handle_.get_stream()); + thrust::gather(handle_.get_thrust_policy(), + vertices_iter, + vertices_iter + vertex_ids.size(), + in_degrees->begin(), + tmp_in_degrees.begin()); + in_degrees = std::move(tmp_in_degrees); + } else { + rmm::device_uvector tmp_out_degrees(vertex_ids.size(), handle_.get_stream()); + thrust::gather(handle_.get_thrust_policy(), + vertices_iter, + vertices_iter + vertex_ids.size(), + out_degrees->begin(), + tmp_out_degrees.begin()); + out_degrees = std::move(tmp_out_degrees); + } + + cugraph::unrenumber_local_int_vertices( + handle_, + vertex_ids.data(), + vertex_ids.size(), + number_map->data(), + graph_view.local_vertex_partition_range_first(), + graph_view.local_vertex_partition_range_last(), + do_expensive_check_); + } else { + vertex_ids.resize(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_degrees_result_t{ + graph_view.is_symmetric(), + new cugraph::c_api::cugraph_type_erased_device_array_t(vertex_ids, graph_->vertex_type_), + in_degrees + ? new cugraph::c_api::cugraph_type_erased_device_array_t(*in_degrees, graph_->edge_type_) + : nullptr, + out_degrees + ? new cugraph::c_api::cugraph_type_erased_device_array_t(*out_degrees, graph_->edge_type_) + : nullptr}; + } + } +}; + +} // namespace + +extern "C" cugraph_error_code_t cugraph_in_degrees( + const cugraph_resource_handle_t* handle, + cugraph_graph_t* graph, + const cugraph_type_erased_device_array_view_t* source_vertices, + bool_t do_expensive_check, + cugraph_degrees_result_t** result, + cugraph_error_t** error) +{ + degrees_functor functor(handle, graph, source_vertices, true, false, do_expensive_check); + + return cugraph::c_api::run_algorithm(graph, functor, result, error); +} + +extern "C" cugraph_error_code_t cugraph_out_degrees( + const cugraph_resource_handle_t* handle, + cugraph_graph_t* graph, + const cugraph_type_erased_device_array_view_t* source_vertices, + bool_t do_expensive_check, + cugraph_degrees_result_t** result, + cugraph_error_t** error) +{ + degrees_functor functor(handle, graph, source_vertices, false, true, do_expensive_check); + + return cugraph::c_api::run_algorithm(graph, functor, result, error); +} + +extern "C" cugraph_error_code_t cugraph_degrees( + const cugraph_resource_handle_t* handle, + cugraph_graph_t* graph, + const cugraph_type_erased_device_array_view_t* source_vertices, + bool_t do_expensive_check, + cugraph_degrees_result_t** result, + cugraph_error_t** error) +{ + degrees_functor functor(handle, graph, source_vertices, true, true, do_expensive_check); + + return cugraph::c_api::run_algorithm(graph, functor, result, error); +} diff --git a/cpp/src/c_api/degrees_result.cpp b/cpp/src/c_api/degrees_result.cpp new file mode 100644 index 00000000000..a4649e36d05 --- /dev/null +++ b/cpp/src/c_api/degrees_result.cpp @@ -0,0 +1,63 @@ +/* + * Copyright (c) 2024, 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_api/degrees_result.hpp" + +#include + +extern "C" cugraph_type_erased_device_array_view_t* cugraph_degrees_result_get_vertices( + cugraph_degrees_result_t* degrees_result) +{ + auto internal_pointer = + reinterpret_cast(degrees_result); + return reinterpret_cast( + internal_pointer->vertex_ids_->view()); +} + +extern "C" cugraph_type_erased_device_array_view_t* cugraph_degrees_result_get_in_degrees( + cugraph_degrees_result_t* degrees_result) +{ + auto internal_pointer = + reinterpret_cast(degrees_result); + return internal_pointer->in_degrees_ == nullptr + ? nullptr + : reinterpret_cast( + internal_pointer->in_degrees_->view()); +} + +extern "C" cugraph_type_erased_device_array_view_t* cugraph_degrees_result_get_out_degrees( + cugraph_degrees_result_t* degrees_result) +{ + auto internal_pointer = + reinterpret_cast(degrees_result); + return internal_pointer->out_degrees_ != nullptr + ? reinterpret_cast( + internal_pointer->out_degrees_->view()) + : internal_pointer->is_symmetric + ? reinterpret_cast( + internal_pointer->in_degrees_->view()) + : nullptr; +} + +extern "C" void cugraph_degrees_result_free(cugraph_degrees_result_t* degrees_result) +{ + auto internal_pointer = + reinterpret_cast(degrees_result); + delete internal_pointer->vertex_ids_; + delete internal_pointer->in_degrees_; + delete internal_pointer->out_degrees_; + delete internal_pointer; +} diff --git a/cpp/src/c_api/degrees_result.hpp b/cpp/src/c_api/degrees_result.hpp new file mode 100644 index 00000000000..c6e9bffa5a1 --- /dev/null +++ b/cpp/src/c_api/degrees_result.hpp @@ -0,0 +1,32 @@ +/* + * Copyright (c) 2024, 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 "c_api/array.hpp" + +namespace cugraph { +namespace c_api { + +struct cugraph_degrees_result_t { + bool is_symmetric{false}; + cugraph_type_erased_device_array_t* vertex_ids_{}; + cugraph_type_erased_device_array_t* in_degrees_{}; + cugraph_type_erased_device_array_t* out_degrees_{}; +}; + +} // namespace c_api +} // namespace cugraph diff --git a/cpp/tests/CMakeLists.txt b/cpp/tests/CMakeLists.txt index af0dffcbf65..c84711e1a69 100644 --- a/cpp/tests/CMakeLists.txt +++ b/cpp/tests/CMakeLists.txt @@ -696,6 +696,7 @@ if(BUILD_CUGRAPH_MG_TESTS) ConfigureCTestMG(MG_CAPI_SIMILARITY_TEST c_api/mg_similarity_test.c) ConfigureCTestMG(MG_CAPI_K_CORE_TEST c_api/mg_k_core_test.c) ConfigureCTestMG(MG_CAPI_INDUCED_SUBGRAPH_TEST c_api/mg_induced_subgraph_test.c) + ConfigureCTestMG(MG_CAPI_DEGREES c_api/mg_degrees_test.c) ConfigureCTestMG(MG_CAPI_EGONET_TEST c_api/mg_egonet_test.c) ConfigureCTestMG(MG_CAPI_TWO_HOP_NEIGHBORS_TEST c_api/mg_two_hop_neighbors_test.c) @@ -764,6 +765,7 @@ ConfigureCTest(CAPI_CORE_NUMBER_TEST c_api/core_number_test.c) ConfigureCTest(CAPI_SIMILARITY_TEST c_api/similarity_test.c) ConfigureCTest(CAPI_K_CORE_TEST c_api/k_core_test.c) ConfigureCTest(CAPI_INDUCED_SUBGRAPH_TEST c_api/induced_subgraph_test.c) +ConfigureCTest(CAPI_DEGREES c_api/degrees_test.c) ConfigureCTest(CAPI_EGONET_TEST c_api/egonet_test.c) ConfigureCTest(CAPI_TWO_HOP_NEIGHBORS_TEST c_api/two_hop_neighbors_test.c) ConfigureCTest(CAPI_LEGACY_K_TRUSS_TEST c_api/legacy_k_truss_test.c) diff --git a/cpp/tests/c_api/c_test_utils.h b/cpp/tests/c_api/c_test_utils.h index ab9fbeccd4b..fbbf6333ee3 100644 --- a/cpp/tests/c_api/c_test_utils.h +++ b/cpp/tests/c_api/c_test_utils.h @@ -1,5 +1,5 @@ /* - * Copyright (c) 2021-2023, NVIDIA CORPORATION. + * Copyright (c) 2021-2024, NVIDIA CORPORATION. * * Licensed under the Apache License, Version 2.0 (the "License"); * you may not use this file except in compliance with the License. @@ -101,6 +101,8 @@ int create_sg_test_graph(const cugraph_resource_handle_t* handle, cugraph_graph_t** graph, cugraph_error_t** ret_error); +size_t cugraph_size_t_allreduce(const cugraph_resource_handle_t* handle, size_t value); + #ifdef __cplusplus } #endif diff --git a/cpp/tests/c_api/degrees_test.c b/cpp/tests/c_api/degrees_test.c new file mode 100644 index 00000000000..10a038b323b --- /dev/null +++ b/cpp/tests/c_api/degrees_test.c @@ -0,0 +1,387 @@ +/* + * Copyright (c) 2024, 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; + +/* + * Simple check of creating a graph from a COO on device memory. + */ +int generic_degrees_test(vertex_t* h_src, + vertex_t* h_dst, + weight_t* h_wgt, + size_t num_vertices, + size_t num_edges, + vertex_t* h_vertices, + size_t num_vertices_to_compute, + bool_t in_degrees, + bool_t out_degrees, + bool_t store_transposed, + bool_t is_symmetric, + edge_t *h_in_degrees, + edge_t *h_out_degrees) +{ + int test_ret_value = 0; + + cugraph_error_code_t ret_code = CUGRAPH_SUCCESS; + cugraph_error_t* ret_error; + + cugraph_resource_handle_t* handle = NULL; + cugraph_graph_t* graph = NULL; + cugraph_degrees_result_t* result = NULL; + + handle = cugraph_create_resource_handle(NULL); + TEST_ASSERT(test_ret_value, handle != NULL, "resource handle creation failed."); + + ret_code = create_test_graph( + handle, h_src, h_dst, h_wgt, num_edges, store_transposed, FALSE, is_symmetric, &graph, &ret_error); + + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "create_test_graph failed."); + TEST_ALWAYS_ASSERT(ret_code == CUGRAPH_SUCCESS, cugraph_error_message(ret_error)); + + if (h_vertices == NULL) { + if (in_degrees && out_degrees) { + ret_code = cugraph_degrees( + handle, graph, NULL, FALSE, &result, &ret_error); + } else if (in_degrees) { + ret_code = cugraph_in_degrees( + handle, graph, NULL, FALSE, &result, &ret_error); + } else { + ret_code = cugraph_out_degrees( + handle, graph, NULL, FALSE, &result, &ret_error); + } + + TEST_ASSERT( + test_ret_value, ret_code == CUGRAPH_SUCCESS, "cugraph_extract_degrees failed."); + } else { + cugraph_type_erased_device_array_t* vertices = NULL; + cugraph_type_erased_device_array_view_t* vertices_view = NULL; + + ret_code = + cugraph_type_erased_device_array_create(handle, num_vertices_to_compute, INT32, &vertices, &ret_error); + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "seeds create failed."); + + vertices_view = cugraph_type_erased_device_array_view(vertices); + + ret_code = cugraph_type_erased_device_array_view_copy_from_host( + handle, vertices_view, (byte_t*)h_vertices, &ret_error); + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "src copy_from_host failed."); + + if (in_degrees && out_degrees) { + ret_code = cugraph_degrees( + handle, graph, vertices_view, FALSE, &result, &ret_error); + } else if (in_degrees) { + ret_code = cugraph_in_degrees( + handle, graph, vertices_view, FALSE, &result, &ret_error); + } else { + ret_code = cugraph_out_degrees( + handle, graph, vertices_view, FALSE, &result, &ret_error); + } + + TEST_ASSERT( + test_ret_value, ret_code == CUGRAPH_SUCCESS, "cugraph_extract_degrees failed."); + } + + cugraph_type_erased_device_array_view_t* result_vertices; + cugraph_type_erased_device_array_view_t* result_in_degrees; + cugraph_type_erased_device_array_view_t* result_out_degrees; + + result_vertices = cugraph_degrees_result_get_vertices(result); + result_in_degrees = cugraph_degrees_result_get_in_degrees(result); + result_out_degrees = cugraph_degrees_result_get_out_degrees(result); + + size_t num_result_vertices = cugraph_type_erased_device_array_view_size(result_vertices); + + vertex_t h_result_vertices[num_result_vertices]; + edge_t h_result_in_degrees[num_result_vertices]; + edge_t h_result_out_degrees[num_result_vertices]; + + ret_code = cugraph_type_erased_device_array_view_copy_to_host( + handle, (byte_t*)h_result_vertices, result_vertices, &ret_error); + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "copy_to_host failed."); + + if (result_in_degrees != NULL) { + ret_code = cugraph_type_erased_device_array_view_copy_to_host( + handle, (byte_t*)h_result_in_degrees, result_in_degrees, &ret_error); + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "copy_to_host failed."); + } + + if (result_out_degrees != NULL) { + ret_code = cugraph_type_erased_device_array_view_copy_to_host( + handle, (byte_t*)h_result_out_degrees, result_out_degrees, &ret_error); + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "copy_to_host failed."); + } + + if (h_vertices != NULL) { + TEST_ASSERT(test_ret_value, num_result_vertices == num_vertices_to_compute, "results not the same size"); + } else { + TEST_ASSERT(test_ret_value, num_result_vertices == num_vertices, "results not the same size"); + } + + for (size_t i = 0; (i < num_result_vertices) && (test_ret_value == 0); ++i) { + if (h_in_degrees != NULL) { + TEST_ASSERT(test_ret_value, h_result_in_degrees[i] == h_in_degrees[h_result_vertices[i]], "in degree did not match"); + } + + if (h_out_degrees != NULL) { + TEST_ASSERT(test_ret_value, h_result_out_degrees[i] == h_out_degrees[h_result_vertices[i]], "out degree did not match"); + } + } + + cugraph_degrees_result_free(result); + cugraph_graph_free(graph); + cugraph_error_free(ret_error); + + return test_ret_value; +} + +int test_degrees() +{ + size_t num_edges = 8; + size_t num_vertices = 6; + + vertex_t h_src[] = {0, 1, 1, 2, 2, 2, 3, 4}; + vertex_t h_dst[] = {1, 3, 4, 0, 1, 3, 5, 5}; + weight_t h_wgt[] = {0.1f, 2.1f, 1.1f, 5.1f, 3.1f, 4.1f, 7.2f, 3.2f}; + vertex_t h_in_degrees[] = {1, 2, 0, 2, 1, 2}; + vertex_t h_out_degrees[] = {1, 2, 3, 1, 1, 0}; + + return generic_degrees_test(h_src, + h_dst, + h_wgt, + num_vertices, + num_edges, + NULL, + 0, + TRUE, + TRUE, + FALSE, + FALSE, + h_in_degrees, + h_out_degrees); +} + +int test_degrees_symmetric() +{ + 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}; + vertex_t h_in_degrees[] = {2, 4, 3, 3, 2, 2}; + vertex_t h_out_degrees[] = {2, 4, 3, 3, 2, 2}; + + return generic_degrees_test(h_src, + h_dst, + h_wgt, + num_vertices, + num_edges, + NULL, + 0, + TRUE, + TRUE, + FALSE, + TRUE, + h_in_degrees, + h_out_degrees); +} + +int test_in_degrees() +{ + size_t num_edges = 8; + size_t num_vertices = 6; + + vertex_t h_src[] = {0, 1, 1, 2, 2, 2, 3, 4}; + vertex_t h_dst[] = {1, 3, 4, 0, 1, 3, 5, 5}; + weight_t h_wgt[] = {0.1f, 2.1f, 1.1f, 5.1f, 3.1f, 4.1f, 7.2f, 3.2f}; + vertex_t h_in_degrees[] = {1, 2, 0, 2, 1, 2}; + + return generic_degrees_test(h_src, + h_dst, + h_wgt, + num_vertices, + num_edges, + NULL, + 0, + TRUE, + FALSE, + FALSE, + TRUE, + h_in_degrees, + NULL); +} + +int test_out_degrees() +{ + size_t num_edges = 8; + size_t num_vertices = 6; + + vertex_t h_src[] = {0, 1, 1, 2, 2, 2, 3, 4}; + vertex_t h_dst[] = {1, 3, 4, 0, 1, 3, 5, 5}; + weight_t h_wgt[] = {0.1f, 2.1f, 1.1f, 5.1f, 3.1f, 4.1f, 7.2f, 3.2f}; + vertex_t h_out_degrees[] = {1, 2, 3, 1, 1, 0}; + + return generic_degrees_test(h_src, + h_dst, + h_wgt, + num_vertices, + num_edges, + NULL, + 0, + FALSE, + TRUE, + FALSE, + TRUE, + NULL, + h_out_degrees); +} + +int test_degrees_subset() +{ + size_t num_edges = 8; + size_t num_vertices = 6; + size_t num_vertices_to_compute = 3; + + vertex_t h_src[] = {0, 1, 1, 2, 2, 2, 3, 4}; + vertex_t h_dst[] = {1, 3, 4, 0, 1, 3, 5, 5}; + weight_t h_wgt[] = {0.1f, 2.1f, 1.1f, 5.1f, 3.1f, 4.1f, 7.2f, 3.2f}; + vertex_t h_vertices[] = {2, 3, 5}; + vertex_t h_in_degrees[] = {-1, -1, 0, 2, -1, 2}; + vertex_t h_out_degrees[] = {-1, -1, 3, 1, -1, 0}; + + return generic_degrees_test(h_src, + h_dst, + h_wgt, + num_vertices, + num_edges, + h_vertices, + num_vertices_to_compute, + TRUE, + TRUE, + FALSE, + FALSE, + h_in_degrees, + h_out_degrees); +} + +int test_degrees_symmetric_subset() +{ + size_t num_edges = 16; + size_t num_vertices = 6; + size_t num_vertices_to_compute = 3; + + 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}; + vertex_t h_vertices[] = {2, 3, 5}; + vertex_t h_in_degrees[] = {-1, -1, 3, 3, -1, 2}; + vertex_t h_out_degrees[] = {-1, -1, 3, 3, -1, 2}; + + return generic_degrees_test(h_src, + h_dst, + h_wgt, + num_vertices, + num_edges, + h_vertices, + num_vertices_to_compute, + TRUE, + TRUE, + FALSE, + TRUE, + h_in_degrees, + h_out_degrees); +} + +int test_in_degrees_subset() +{ + size_t num_edges = 8; + size_t num_vertices = 6; + size_t num_vertices_to_compute = 3; + + vertex_t h_src[] = {0, 1, 1, 2, 2, 2, 3, 4}; + vertex_t h_dst[] = {1, 3, 4, 0, 1, 3, 5, 5}; + weight_t h_wgt[] = {0.1f, 2.1f, 1.1f, 5.1f, 3.1f, 4.1f, 7.2f, 3.2f}; + vertex_t h_vertices[] = {2, 3, 5}; + vertex_t h_in_degrees[] = {-1, -1, 0, 2, -1, 2}; + + return generic_degrees_test(h_src, + h_dst, + h_wgt, + num_vertices, + num_edges, + h_vertices, + num_vertices_to_compute, + TRUE, + FALSE, + FALSE, + TRUE, + h_in_degrees, + NULL); +} + +int test_out_degrees_subset() +{ + size_t num_edges = 8; + size_t num_vertices = 6; + size_t num_vertices_to_compute = 3; + + vertex_t h_src[] = {0, 1, 1, 2, 2, 2, 3, 4}; + vertex_t h_dst[] = {1, 3, 4, 0, 1, 3, 5, 5}; + weight_t h_wgt[] = {0.1f, 2.1f, 1.1f, 5.1f, 3.1f, 4.1f, 7.2f, 3.2f}; + vertex_t h_vertices[] = {2, 3, 5}; + vertex_t h_out_degrees[] = {-1, -1, 3, 1, -1, 0}; + + return generic_degrees_test(h_src, + h_dst, + h_wgt, + num_vertices, + num_edges, + h_vertices, + num_vertices_to_compute, + FALSE, + TRUE, + FALSE, + TRUE, + NULL, + h_out_degrees); +} + +/******************************************************************************/ + +int main(int argc, char** argv) +{ + int result = 0; + result |= RUN_TEST(test_degrees); + result |= RUN_TEST(test_degrees_symmetric); + result |= RUN_TEST(test_in_degrees); + result |= RUN_TEST(test_out_degrees); + result |= RUN_TEST(test_degrees_subset); + result |= RUN_TEST(test_degrees_symmetric_subset); + result |= RUN_TEST(test_in_degrees_subset); + result |= RUN_TEST(test_out_degrees_subset); + return result; +} diff --git a/cpp/tests/c_api/mg_degrees_test.c b/cpp/tests/c_api/mg_degrees_test.c new file mode 100644 index 00000000000..3312dd4f5bb --- /dev/null +++ b/cpp/tests/c_api/mg_degrees_test.c @@ -0,0 +1,407 @@ +/* + * Copyright (c) 2024, 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; + +/* + * Simple check of creating a graph from a COO on device memory. + */ +int generic_degrees_test(const cugraph_resource_handle_t* handle, + vertex_t* h_src, + vertex_t* h_dst, + weight_t* h_wgt, + size_t num_vertices, + size_t num_edges, + vertex_t* h_vertices, + size_t num_vertices_to_compute, + bool_t in_degrees, + bool_t out_degrees, + bool_t store_transposed, + bool_t is_symmetric, + edge_t* h_in_degrees, + edge_t* h_out_degrees) +{ + int test_ret_value = 0; + + cugraph_error_code_t ret_code = CUGRAPH_SUCCESS; + cugraph_error_t* ret_error; + + cugraph_graph_t* graph = NULL; + cugraph_degrees_result_t* result = NULL; + + ret_code = create_mg_test_graph( + handle, h_src, h_dst, h_wgt, num_edges, store_transposed, is_symmetric, &graph, &ret_error); + + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "create_test_graph failed."); + TEST_ALWAYS_ASSERT(ret_code == CUGRAPH_SUCCESS, cugraph_error_message(ret_error)); + + if (h_vertices == NULL) { + if (in_degrees && out_degrees) { + ret_code = cugraph_degrees( + handle, graph, NULL, FALSE, &result, &ret_error); + } else if (in_degrees) { + ret_code = cugraph_in_degrees( + handle, graph, NULL, FALSE, &result, &ret_error); + } else { + ret_code = cugraph_out_degrees( + handle, graph, NULL, FALSE, &result, &ret_error); + } + + TEST_ASSERT( + test_ret_value, ret_code == CUGRAPH_SUCCESS, "cugraph_extract_degrees failed."); + } else { + cugraph_type_erased_device_array_t* vertices = NULL; + cugraph_type_erased_device_array_view_t* vertices_view = NULL; + + int rank = cugraph_resource_handle_get_rank(handle); + + size_t num_to_allocate = 0; + if (rank == 0) num_to_allocate = num_vertices_to_compute; + + ret_code = + cugraph_type_erased_device_array_create(handle, num_to_allocate, INT32, &vertices, &ret_error); + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "seeds create failed."); + + vertices_view = cugraph_type_erased_device_array_view(vertices); + + ret_code = cugraph_type_erased_device_array_view_copy_from_host( + handle, vertices_view, (byte_t*)h_vertices, &ret_error); + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "src copy_from_host failed."); + + if (in_degrees && out_degrees) { + ret_code = cugraph_degrees( + handle, graph, vertices_view, FALSE, &result, &ret_error); + } else if (in_degrees) { + ret_code = cugraph_in_degrees( + handle, graph, vertices_view, FALSE, &result, &ret_error); + } else { + ret_code = cugraph_out_degrees( + handle, graph, vertices_view, FALSE, &result, &ret_error); + } + + TEST_ASSERT( + test_ret_value, ret_code == CUGRAPH_SUCCESS, "cugraph_extract_degrees failed."); + } + + cugraph_type_erased_device_array_view_t* result_vertices; + cugraph_type_erased_device_array_view_t* result_in_degrees; + cugraph_type_erased_device_array_view_t* result_out_degrees; + + result_vertices = cugraph_degrees_result_get_vertices(result); + result_in_degrees = cugraph_degrees_result_get_in_degrees(result); + result_out_degrees = cugraph_degrees_result_get_out_degrees(result); + + size_t num_result_vertices = cugraph_type_erased_device_array_view_size(result_vertices); + + vertex_t h_result_vertices[num_result_vertices]; + edge_t h_result_in_degrees[num_result_vertices]; + edge_t h_result_out_degrees[num_result_vertices]; + + ret_code = cugraph_type_erased_device_array_view_copy_to_host( + handle, (byte_t*)h_result_vertices, result_vertices, &ret_error); + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "copy_to_host failed."); + + if (result_in_degrees != NULL) { + ret_code = cugraph_type_erased_device_array_view_copy_to_host( + handle, (byte_t*)h_result_in_degrees, result_in_degrees, &ret_error); + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "copy_to_host failed."); + } + + if (result_out_degrees != NULL) { + ret_code = cugraph_type_erased_device_array_view_copy_to_host( + handle, (byte_t*)h_result_out_degrees, result_out_degrees, &ret_error); + TEST_ASSERT(test_ret_value, ret_code == CUGRAPH_SUCCESS, "copy_to_host failed."); + } + + if (h_vertices != NULL) { + size_t xxx = cugraph_size_t_allreduce(handle, num_result_vertices); + TEST_ASSERT(test_ret_value, cugraph_size_t_allreduce(handle, num_result_vertices) == num_vertices_to_compute, "results not the same size"); + } else { + size_t xxx = cugraph_size_t_allreduce(handle, num_result_vertices); + TEST_ASSERT(test_ret_value, cugraph_size_t_allreduce(handle, num_result_vertices) == num_vertices, "results not the same size"); + } + + for (size_t i = 0; (i < num_result_vertices) && (test_ret_value == 0); ++i) { + if (h_in_degrees != NULL) { + TEST_ASSERT(test_ret_value, h_result_in_degrees[i] == h_in_degrees[h_result_vertices[i]], "in degree did not match"); + } + + if (h_out_degrees != NULL) { + TEST_ASSERT(test_ret_value, h_result_out_degrees[i] == h_out_degrees[h_result_vertices[i]], "out degree did not match"); + } + } + + cugraph_degrees_result_free(result); + cugraph_graph_free(graph); + cugraph_error_free(ret_error); + return test_ret_value; +} + +int test_degrees(const cugraph_resource_handle_t* handle) +{ + size_t num_edges = 8; + size_t num_vertices = 6; + + vertex_t h_src[] = {0, 1, 1, 2, 2, 2, 3, 4}; + vertex_t h_dst[] = {1, 3, 4, 0, 1, 3, 5, 5}; + weight_t h_wgt[] = {0.1f, 2.1f, 1.1f, 5.1f, 3.1f, 4.1f, 7.2f, 3.2f}; + vertex_t h_in_degrees[] = {1, 2, 0, 2, 1, 2}; + vertex_t h_out_degrees[] = {1, 2, 3, 1, 1, 0}; + + // Pagerank wants store_transposed = TRUE + return generic_degrees_test(handle, + h_src, + h_dst, + h_wgt, + num_vertices, + num_edges, + NULL, + 0, + TRUE, + TRUE, + TRUE, + FALSE, + h_in_degrees, + h_out_degrees); +} + +int test_degrees_symmetric(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}; + vertex_t h_in_degrees[] = {2, 4, 3, 3, 2, 2}; + vertex_t h_out_degrees[] = {2, 4, 3, 3, 2, 2}; + + // Pagerank wants store_transposed = TRUE + return generic_degrees_test(handle, + h_src, + h_dst, + h_wgt, + num_vertices, + num_edges, + NULL, + 0, + TRUE, + TRUE, + TRUE, + TRUE, + h_in_degrees, + h_out_degrees); +} + +int test_in_degrees(const cugraph_resource_handle_t *handle) +{ + size_t num_edges = 8; + size_t num_vertices = 6; + + vertex_t h_src[] = {0, 1, 1, 2, 2, 2, 3, 4}; + vertex_t h_dst[] = {1, 3, 4, 0, 1, 3, 5, 5}; + weight_t h_wgt[] = {0.1f, 2.1f, 1.1f, 5.1f, 3.1f, 4.1f, 7.2f, 3.2f}; + vertex_t h_in_degrees[] = {1, 2, 0, 2, 1, 2}; + + return generic_degrees_test(handle, + h_src, + h_dst, + h_wgt, + num_vertices, + num_edges, + NULL, + 0, + TRUE, + FALSE, + FALSE, + TRUE, + h_in_degrees, + NULL); +} + +int test_out_degrees(const cugraph_resource_handle_t *handle) +{ + size_t num_edges = 8; + size_t num_vertices = 6; + + vertex_t h_src[] = {0, 1, 1, 2, 2, 2, 3, 4}; + vertex_t h_dst[] = {1, 3, 4, 0, 1, 3, 5, 5}; + weight_t h_wgt[] = {0.1f, 2.1f, 1.1f, 5.1f, 3.1f, 4.1f, 7.2f, 3.2f}; + vertex_t h_out_degrees[] = {1, 2, 3, 1, 1, 0}; + + return generic_degrees_test(handle, + h_src, + h_dst, + h_wgt, + num_vertices, + num_edges, + NULL, + 0, + FALSE, + TRUE, + FALSE, + TRUE, + NULL, + h_out_degrees); +} + +int test_degrees_subset(const cugraph_resource_handle_t* handle) +{ + size_t num_edges = 8; + size_t num_vertices = 6; + size_t num_vertices_to_compute = 3; + + vertex_t h_src[] = {0, 1, 1, 2, 2, 2, 3, 4}; + vertex_t h_dst[] = {1, 3, 4, 0, 1, 3, 5, 5}; + weight_t h_wgt[] = {0.1f, 2.1f, 1.1f, 5.1f, 3.1f, 4.1f, 7.2f, 3.2f}; + vertex_t h_vertices[] = {2, 3, 5}; + vertex_t h_in_degrees[] = {-1, -1, 0, 2, -1, 2}; + vertex_t h_out_degrees[] = {-1, -1, 3, 1, -1, 0}; + + return generic_degrees_test(handle, + h_src, + h_dst, + h_wgt, + num_vertices, + num_edges, + h_vertices, + num_vertices_to_compute, + TRUE, + TRUE, + FALSE, + FALSE, + h_in_degrees, + h_out_degrees); +} + +int test_degrees_symmetric_subset(const cugraph_resource_handle_t* handle) +{ + size_t num_edges = 16; + size_t num_vertices = 6; + size_t num_vertices_to_compute = 3; + + 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}; + vertex_t h_vertices[] = {2, 3, 5}; + vertex_t h_in_degrees[] = {-1, -1, 3, 3, -1, 2}; + vertex_t h_out_degrees[] = {-1, -1, 3, 3, -1, 2}; + + return generic_degrees_test(handle, + h_src, + h_dst, + h_wgt, + num_vertices, + num_edges, + h_vertices, + num_vertices_to_compute, + TRUE, + TRUE, + FALSE, + TRUE, + h_in_degrees, + h_out_degrees); +} + +int test_in_degrees_subset(const cugraph_resource_handle_t* handle) +{ + size_t num_edges = 8; + size_t num_vertices = 6; + size_t num_vertices_to_compute = 3; + + vertex_t h_src[] = {0, 1, 1, 2, 2, 2, 3, 4}; + vertex_t h_dst[] = {1, 3, 4, 0, 1, 3, 5, 5}; + weight_t h_wgt[] = {0.1f, 2.1f, 1.1f, 5.1f, 3.1f, 4.1f, 7.2f, 3.2f}; + vertex_t h_vertices[] = {2, 3, 5}; + vertex_t h_in_degrees[] = {-1, -1, 0, 2, -1, 2}; + + return generic_degrees_test(handle, + h_src, + h_dst, + h_wgt, + num_vertices, + num_edges, + h_vertices, + num_vertices_to_compute, + TRUE, + FALSE, + FALSE, + TRUE, + h_in_degrees, + NULL); +} + +int test_out_degrees_subset(const cugraph_resource_handle_t* handle) +{ + size_t num_edges = 8; + size_t num_vertices = 6; + size_t num_vertices_to_compute = 3; + + vertex_t h_src[] = {0, 1, 1, 2, 2, 2, 3, 4}; + vertex_t h_dst[] = {1, 3, 4, 0, 1, 3, 5, 5}; + weight_t h_wgt[] = {0.1f, 2.1f, 1.1f, 5.1f, 3.1f, 4.1f, 7.2f, 3.2f}; + vertex_t h_vertices[] = {2, 3, 5}; + vertex_t h_out_degrees[] = {-1, -1, 3, 1, -1, 0}; + + return generic_degrees_test(handle, + h_src, + h_dst, + h_wgt, + num_vertices, + num_edges, + h_vertices, + num_vertices_to_compute, + FALSE, + TRUE, + FALSE, + TRUE, + NULL, + h_out_degrees); +} + +/******************************************************************************/ + +int main(int argc, char** argv) +{ + void* raft_handle = create_mg_raft_handle(argc, argv); + cugraph_resource_handle_t* handle = cugraph_create_resource_handle(raft_handle); + + int result = 0; + result |= RUN_MG_TEST(test_degrees, handle); + result |= RUN_MG_TEST(test_degrees_symmetric, handle); + result |= RUN_MG_TEST(test_in_degrees, handle); + result |= RUN_MG_TEST(test_out_degrees, handle); + result |= RUN_MG_TEST(test_degrees_subset, handle); + result |= RUN_MG_TEST(test_degrees_symmetric_subset, handle); + result |= RUN_MG_TEST(test_in_degrees_subset, handle); + result |= RUN_MG_TEST(test_out_degrees_subset, handle); + + cugraph_free_resource_handle(handle); + free_mg_raft_handle(raft_handle); + + return result; +} diff --git a/cpp/tests/c_api/test_utils.cpp b/cpp/tests/c_api/test_utils.cpp index e37cc4555dd..3013cbb7cc6 100644 --- a/cpp/tests/c_api/test_utils.cpp +++ b/cpp/tests/c_api/test_utils.cpp @@ -1,5 +1,5 @@ /* - * Copyright (c) 2021-2023, NVIDIA CORPORATION. + * Copyright (c) 2021-2024, NVIDIA CORPORATION. * * Licensed under the Apache License, Version 2.0 (the "License"); * you may not use this file except in compliance with the License. @@ -15,6 +15,9 @@ */ #include "c_test_utils.h" +#include "c_api/resource_handle.hpp" + +#include #include @@ -388,3 +391,12 @@ int create_sg_test_graph(const cugraph_resource_handle_t* handle, return test_ret_value; } + +extern "C" size_t cugraph_size_t_allreduce(const cugraph_resource_handle_t* handle, size_t value) +{ + auto internal_handle = reinterpret_cast(handle); + return cugraph::host_scalar_allreduce(internal_handle->handle_->get_comms(), + value, + raft::comms::op_t::SUM, + internal_handle->handle_->get_stream()); +} diff --git a/python/cugraph/cugraph/structure/graph_implementation/simpleDistributedGraph.py b/python/cugraph/cugraph/structure/graph_implementation/simpleDistributedGraph.py index cdf1e937e67..0ef5eaf1b9e 100644 --- a/python/cugraph/cugraph/structure/graph_implementation/simpleDistributedGraph.py +++ b/python/cugraph/cugraph/structure/graph_implementation/simpleDistributedGraph.py @@ -12,7 +12,7 @@ # limitations under the License. import gc -from typing import Union +from typing import Union, Iterable import warnings import cudf @@ -28,10 +28,11 @@ GraphProperties, get_two_hop_neighbors as pylibcugraph_get_two_hop_neighbors, select_random_vertices as pylibcugraph_select_random_vertices, + degrees as pylibcugraph_degrees, + in_degrees as pylibcugraph_in_degrees, + out_degrees as pylibcugraph_out_degrees, ) -from cugraph.structure import graph_primtypes_wrapper -from cugraph.structure.graph_primtypes_wrapper import Direction from cugraph.structure.number_map import NumberMap from cugraph.structure.symmetrize import symmetrize from cugraph.dask.common.part_utils import ( @@ -536,7 +537,158 @@ def number_of_edges(self, directed_edges=False): raise RuntimeError("Graph is Empty") return self.properties.edge_count - def in_degree(self, vertex_subset=None): + def degrees_function( + self, + vertex_subset: Union[cudf.Series, dask_cudf.Series, Iterable] = None, + degree_type: str = "in_degree", + ) -> dask_cudf.DataFrame: + """ + Compute vertex in-degree, out-degree, degree and degrees. + + 1) Vertex in-degree is the number of edges pointing into the vertex. + 2) Vertex out-degree is the number of edges pointing out from the vertex. + 3) Vertex degree, is the total number of edges incident to a vertex + (both in and out edges) + 4) Vertex degrees computes vertex in-degree and out-degree. + + By default, this method computes vertex in-degree, out-degree, degree + or degrees for the entire set of vertices. If vertex_subset is provided, + this method optionally filters out all but those listed in + vertex_subset. + + Parameters + ---------- + vertex_subset : cudf.Series or dask_cudf.Series, iterable container, optional + A container of vertices for displaying corresponding in-degree. + If not set, degrees are computed for the entire set of vertices. + + degree_type : str (default='in_degree') + + Returns + ------- + df : dask_cudf.DataFrame + GPU DataFrame of size N (the default) or the size of the given + vertices (vertex_subset) containing the in_degree, out_degrees, + degree or degrees. The ordering is relative to the adjacency list, + or that given by the specified vertex_subset. + + Examples + -------- + >>> M = cudf.read_csv(datasets_path / 'karate.csv', delimiter=' ', + ... dtype=['int32', 'int32', 'float32'], header=None) + >>> G = cugraph.Graph() + >>> G.from_cudf_edgelist(M, '0', '1') + >>> df = G.degrees_function([0,9,12], "in_degree") + + """ + _client = default_client() + + def _call_plc_degrees_function( + sID: bytes, mg_graph_x, source_vertices: cudf.Series, degree_type: str + ) -> cp.array: + + if degree_type == "in_degree": + results = pylibcugraph_in_degrees( + resource_handle=ResourceHandle(Comms.get_handle(sID).getHandle()), + graph=mg_graph_x, + source_vertices=source_vertices, + do_expensive_check=False, + ) + elif degree_type == "out_degree": + results = pylibcugraph_out_degrees( + resource_handle=ResourceHandle(Comms.get_handle(sID).getHandle()), + graph=mg_graph_x, + source_vertices=source_vertices, + do_expensive_check=False, + ) + elif degree_type in ["degree", "degrees"]: + results = pylibcugraph_degrees( + resource_handle=ResourceHandle(Comms.get_handle(sID).getHandle()), + graph=mg_graph_x, + source_vertices=source_vertices, + do_expensive_check=False, + ) + else: + raise ValueError( + "Incorrect degree type passed, valid values are ", + "'in_degree', 'out_degree', 'degree' and 'degrees' ", + f"got '{degree_type}'", + ) + + return results + + if isinstance(vertex_subset, int): + vertex_subset = [vertex_subset] + + if isinstance(vertex_subset, list): + vertex_subset = cudf.Series(vertex_subset) + + if vertex_subset is not None: + if self.renumbered: + vertex_subset = self.renumber_map.to_internal_vertex_id(vertex_subset) + vertex_subset_type = self.edgelist.edgelist_df.dtypes.iloc[0] + else: + vertex_subset_type = self.input_df.dtypes.iloc[0] + + vertex_subset = vertex_subset.astype(vertex_subset_type) + + cupy_result = [ + _client.submit( + _call_plc_degrees_function, + Comms.get_session_id(), + self._plc_graph[w], + vertex_subset, + degree_type, + workers=[w], + allow_other_workers=False, + ) + for w in Comms.get_workers() + ] + + wait(cupy_result) + + def convert_to_cudf(cp_arrays: cp.ndarray, degree_type: bool) -> cudf.DataFrame: + """ + Creates a cudf DataFrame from cupy arrays from pylibcugraph wrapper + """ + df = cudf.DataFrame() + df["vertex"] = cp_arrays[0] + if degree_type in ["in_degree", "out_degree"]: + df["degree"] = cp_arrays[1] + # degree_type must be either 'degree' or 'degrees' + else: + if degree_type == "degrees": + df["in_degree"] = cp_arrays[1] + df["out_degree"] = cp_arrays[2] + else: + df["degree"] = cp_arrays[1] + cp_arrays[2] + return df + + cudf_result = [ + _client.submit( + convert_to_cudf, + cp_arrays, + degree_type, + workers=_client.who_has(cp_arrays)[cp_arrays.key], + ) + for cp_arrays in cupy_result + ] + + wait(cudf_result) + ddf = dask_cudf.from_delayed(cudf_result).persist() + wait(ddf) + + # Wait until the inactive futures are released + wait([(r.release(), c_r.release()) for r, c_r in zip(cupy_result, cudf_result)]) + + if self.properties.renumbered: + ddf = self.renumber_map.unrenumber(ddf, "vertex") + + return ddf + + def in_degree( + self, vertex_subset: Union[cudf.Series, dask_cudf.Series, Iterable] = None + ) -> dask_cudf.DataFrame: """ Compute vertex in-degree. Vertex in-degree is the number of edges pointing into the vertex. By default, this method computes vertex @@ -572,61 +724,11 @@ def in_degree(self, vertex_subset=None): >>> df = G.in_degree([0,9,12]) """ - src_col_name = self.source_columns - dst_col_name = self.destination_columns - - # select only the vertex columns - if not isinstance(src_col_name, list) and not isinstance(dst_col_name, list): - vertex_col_names = [src_col_name] + [dst_col_name] - - df = self.input_df[vertex_col_names] - df = df.drop(columns=src_col_name) - - nodes = self.nodes() - if isinstance(nodes, dask_cudf.Series): - nodes = nodes.to_frame() - - if not isinstance(dst_col_name, list): - df = df.rename(columns={dst_col_name: "vertex"}) - dst_col_name = "vertex" - - vertex_col_names = df.columns - nodes.columns = vertex_col_names - - df["degree"] = 1 - - # FIXME: leverage the C++ in_degree for optimal performance - in_degree = ( - df.groupby(dst_col_name) - .degree.count(split_out=df.npartitions) - .reset_index() - ) - - # Add vertices with zero in_degree - in_degree = nodes.merge(in_degree, how="outer").fillna(0) - - # Convert vertex_subset to dataframe. - if vertex_subset is not None: - if not isinstance(vertex_subset, (dask_cudf.DataFrame, cudf.DataFrame)): - if isinstance(vertex_subset, dask_cudf.Series): - vertex_subset = vertex_subset.to_frame() - else: - df = cudf.DataFrame() - if isinstance(vertex_subset, (cudf.Series, list)): - df["vertex"] = vertex_subset - vertex_subset = df - if isinstance(vertex_subset, (dask_cudf.DataFrame, cudf.DataFrame)): - vertex_subset.columns = vertex_col_names - in_degree = in_degree.merge(vertex_subset, how="inner") - else: - raise TypeError( - f"Expected type are: cudf, dask_cudf objects, " - f"iterable container, got " - f"{type(vertex_subset)}" - ) - return in_degree + return self.degrees_function(vertex_subset, "in_degree") - def out_degree(self, vertex_subset=None): + def out_degree( + self, vertex_subset: Union[cudf.Series, dask_cudf.Series, Iterable] = None + ) -> dask_cudf.DataFrame: """ Compute vertex out-degree. Vertex out-degree is the number of edges pointing out from the vertex. By default, this method computes vertex @@ -662,62 +764,11 @@ def out_degree(self, vertex_subset=None): >>> df = G.out_degree([0,9,12]) """ - src_col_name = self.source_columns - dst_col_name = self.destination_columns - - # select only the vertex columns - if not isinstance(src_col_name, list) and not isinstance(dst_col_name, list): - vertex_col_names = [src_col_name] + [dst_col_name] - - df = self.input_df[vertex_col_names] - df = df.drop(columns=dst_col_name) - - nodes = self.nodes() - if isinstance(nodes, dask_cudf.Series): - nodes = nodes.to_frame() - - if not isinstance(src_col_name, list): - df = df.rename(columns={src_col_name: "vertex"}) - src_col_name = "vertex" - - vertex_col_names = df.columns - - nodes.columns = vertex_col_names - - df["degree"] = 1 - # leverage the C++ out_degree for optimal performance - out_degree = ( - df.groupby(src_col_name) - .degree.count(split_out=df.npartitions) - .reset_index() - ) - - # Add vertices with zero out_degree - out_degree = nodes.merge(out_degree, how="outer").fillna(0) - - # Convert vertex_subset to dataframe. - if vertex_subset is not None: - if not isinstance(vertex_subset, (dask_cudf.DataFrame, cudf.DataFrame)): - if isinstance(vertex_subset, dask_cudf.Series): - vertex_subset = vertex_subset.to_frame() - else: - df = cudf.DataFrame() - if isinstance(vertex_subset, (cudf.Series, list)): - df["vertex"] = vertex_subset - vertex_subset = df - if isinstance(vertex_subset, (dask_cudf.DataFrame, cudf.DataFrame)): - vertex_subset.columns = vertex_col_names - out_degree = out_degree.merge(vertex_subset, how="inner") - else: - raise TypeError( - f"Expected type are: cudf, dask_cudf objects, " - f"iterable container, got " - f"{type(vertex_subset)}" - ) + return self.degrees_function(vertex_subset, "out_degree") - return out_degree - - def degree(self, vertex_subset=None): + def degree( + self, vertex_subset: Union[cudf.Series, dask_cudf.Series, Iterable] = None + ) -> dask_cudf.DataFrame: """ Compute vertex degree, which is the total number of edges incident to a vertex (both in and out edges). By default, this method computes @@ -754,18 +805,12 @@ def degree(self, vertex_subset=None): """ - vertex_in_degree = self.in_degree(vertex_subset) - vertex_out_degree = self.out_degree(vertex_subset) - # FIXME: leverage the C++ degree for optimal performance - vertex_degree = dask_cudf.concat([vertex_in_degree, vertex_out_degree]) - vertex_degree = vertex_degree.groupby(["vertex"], as_index=False).sum( - split_out=self.input_df.npartitions - ) - - return vertex_degree + return self.degrees_function(vertex_subset, "degree") # FIXME: vertex_subset could be a DataFrame for multi-column vertices - def degrees(self, vertex_subset=None): + def degrees( + self, vertex_subset: Union[cudf.Series, dask_cudf.Series, Iterable] = None + ) -> dask_cudf.DataFrame: """ Compute vertex in-degree and out-degree. By default, this method computes vertex degrees for the entire set of vertices. If @@ -802,21 +847,7 @@ def degrees(self, vertex_subset=None): >>> df = G.degrees([0,9,12]) """ - raise NotImplementedError("Not supported for distributed graph") - - def _degree(self, vertex_subset, direction=Direction.ALL): - vertex_col, degree_col = graph_primtypes_wrapper._mg_degree(self, direction) - df = cudf.DataFrame() - df["vertex"] = vertex_col - df["degree"] = degree_col - - if self.renumbered is True: - df = self.renumber_map.unrenumber(df, "vertex") - - if vertex_subset is not None: - df = df[df["vertex"].isin(vertex_subset)] - - return df + return self.degrees_function(vertex_subset, "degrees") def get_two_hop_neighbors(self, start_vertices=None): """ diff --git a/python/cugraph/cugraph/structure/graph_implementation/simpleGraph.py b/python/cugraph/cugraph/structure/graph_implementation/simpleGraph.py index 121a4c6245a..99934e02b10 100644 --- a/python/cugraph/cugraph/structure/graph_implementation/simpleGraph.py +++ b/python/cugraph/cugraph/structure/graph_implementation/simpleGraph.py @@ -12,7 +12,6 @@ # limitations under the License. from cugraph.structure import graph_primtypes_wrapper -from cugraph.structure.graph_primtypes_wrapper import Direction from cugraph.structure.symmetrize import symmetrize from cugraph.structure.number_map import NumberMap import cugraph.dask.common.mg_utils as mg_utils @@ -23,10 +22,13 @@ import numpy as np import warnings from cugraph.dask.structure import replication -from typing import Union, Dict +from typing import Union, Dict, Iterable from pylibcugraph import ( get_two_hop_neighbors as pylibcugraph_get_two_hop_neighbors, select_random_vertices as pylibcugraph_select_random_vertices, + degrees as pylibcugraph_degrees, + in_degrees as pylibcugraph_in_degrees, + out_degrees as pylibcugraph_out_degrees, ) from pylibcugraph import ( @@ -854,7 +856,111 @@ def number_of_edges(self, directed_edges=False): raise ValueError("Graph is Empty") return self.properties.edge_count - def in_degree(self, vertex_subset=None): + def degrees_function( + self, + vertex_subset: Union[cudf.Series, Iterable] = None, + degree_type: str = "in_degree", + ) -> cudf.DataFrame: + """ + Compute vertex in-degree, out-degree, degree and degrees. + + 1) Vertex in-degree is the number of edges pointing into the vertex. + 2) Vertex out-degree is the number of edges pointing out from the vertex. + 3) Vertex degree, is the total number of edges incident to a vertex + (both in and out edges) + 4) Vertex degrees computes vertex in-degree and out-degree. + + By default, this method computes vertex in-degree, out-degree, degree + or degrees for the entire set of vertices. If vertex_subset is provided, + this method optionally filters out all but those listed in + vertex_subset. + + Parameters + ---------- + vertex_subset : cudf.Series or iterable container, optional + A container of vertices for displaying corresponding in-degree. + If not set, degrees are computed for the entire set of vertices. + + degree_type : str (default='in_degree') + + Returns + ------- + df : cudf.DataFrame + GPU DataFrame of size N (the default) or the size of the given + vertices (vertex_subset) containing the in_degree, out_degrees, + degree or degrees. The ordering is relative to the adjacency list, + or that given by the specified vertex_subset. + + Examples + -------- + >>> M = cudf.read_csv(datasets_path / 'karate.csv', delimiter=' ', + ... dtype=['int32', 'int32', 'float32'], header=None) + >>> G = cugraph.Graph() + >>> G.from_cudf_edgelist(M, '0', '1') + >>> df = G.degrees_function([0,9,12], "in_degree") + + """ + if vertex_subset is not None: + if not isinstance(vertex_subset, cudf.Series): + vertex_subset = cudf.Series(vertex_subset) + if self.properties.renumbered is True: + vertex_subset = self.renumber_map.to_internal_vertex_id( + vertex_subset + ) + vertex_subset_type = self.edgelist.edgelist_df.dtypes.iloc[0] + else: + vertex_subset_type = self.input_df.dtypes.iloc[0] + + vertex_subset = vertex_subset.astype(vertex_subset_type) + + do_expensive_check = False + df = cudf.DataFrame() + vertex = None + + if degree_type == "in_degree": + vertex, in_degrees = pylibcugraph_in_degrees( + resource_handle=ResourceHandle(), + graph=self._plc_graph, + source_vertices=vertex_subset, + do_expensive_check=do_expensive_check, + ) + df["degree"] = in_degrees + elif degree_type == "out_degree": + vertex, out_degrees = pylibcugraph_out_degrees( + resource_handle=ResourceHandle(), + graph=self._plc_graph, + source_vertices=vertex_subset, + do_expensive_check=do_expensive_check, + ) + df["degree"] = out_degrees + elif degree_type in ["degree", "degrees"]: + vertex, in_degrees, out_degrees = pylibcugraph_degrees( + resource_handle=ResourceHandle(), + graph=self._plc_graph, + source_vertices=vertex_subset, + do_expensive_check=do_expensive_check, + ) + if degree_type == "degrees": + df["in_degree"] = in_degrees + df["out_degree"] = out_degrees + + else: + df["degree"] = in_degrees + out_degrees + else: + raise ValueError( + "Incorrect degree type passed, valid values are ", + "'in_degree', 'out_degree', 'degree' and 'degrees' ", + f"got '{degree_type}'", + ) + df["vertex"] = vertex + if self.properties.renumbered is True: + df = self.renumber_map.unrenumber(df, "vertex") + + return df + + def in_degree( + self, vertex_subset: Union[cudf.Series, Iterable] = None + ) -> cudf.DataFrame: """ Compute vertex in-degree. Vertex in-degree is the number of edges pointing into the vertex. By default, this method computes vertex @@ -892,11 +998,11 @@ def in_degree(self, vertex_subset=None): >>> df = G.in_degree([0,9,12]) """ - in_degree = self._degree(vertex_subset, direction=Direction.IN) - - return in_degree + return self.degrees_function(vertex_subset, "in_degree") - def out_degree(self, vertex_subset=None): + def out_degree( + self, vertex_subset: Union[cudf.Series, Iterable] = None + ) -> cudf.DataFrame: """ Compute vertex out-degree. Vertex out-degree is the number of edges pointing out from the vertex. By default, this method computes vertex @@ -934,10 +1040,11 @@ def out_degree(self, vertex_subset=None): >>> df = G.out_degree([0,9,12]) """ - out_degree = self._degree(vertex_subset, direction=Direction.OUT) - return out_degree + return self.degrees_function(vertex_subset, "out_degree") - def degree(self, vertex_subset=None): + def degree( + self, vertex_subset: Union[cudf.Series, Iterable] = None + ) -> cudf.DataFrame: """ Compute vertex degree, which is the total number of edges incident to a vertex (both in and out edges). By default, this method computes @@ -976,10 +1083,12 @@ def degree(self, vertex_subset=None): >>> subset_df = G.degree([0,9,12]) """ - return self._degree(vertex_subset) + return self.degrees_function(vertex_subset, "degree") # FIXME: vertex_subset could be a DataFrame for multi-column vertices - def degrees(self, vertex_subset=None): + def degrees( + self, vertex_subset: Union[cudf.Series, Iterable] = None + ) -> cudf.DataFrame: """ Compute vertex in-degree and out-degree. By default, this method computes vertex degrees for the entire set of vertices. If @@ -1019,70 +1128,7 @@ def degrees(self, vertex_subset=None): >>> df = G.degrees([0,9,12]) """ - ( - vertex_col, - in_degree_col, - out_degree_col, - ) = graph_primtypes_wrapper._degrees(self) - - df = cudf.DataFrame() - df["vertex"] = vertex_col - df["in_degree"] = in_degree_col - df["out_degree"] = out_degree_col - - if self.properties.renumbered: - # Get the internal vertex IDs - nodes = self.renumber_map.df_internal_to_external["id"] - else: - nodes = self.nodes() - # If the vertex IDs are not contiguous, remove results for the - # isolated vertices - df = df[df["vertex"].isin(nodes.to_cupy())] - - if vertex_subset is not None: - if not isinstance(vertex_subset, cudf.Series): - vertex_subset = cudf.Series(vertex_subset) - if self.properties.renumbered: - vertex_subset = self.renumber_map.to_internal_vertex_id( - vertex_subset - ) - vertex_subset = vertex_subset.to_cupy() - df = df[df["vertex"].isin(vertex_subset)] - - if self.properties.renumbered: - df = self.renumber_map.unrenumber(df, "vertex") - - return df - - def _degree(self, vertex_subset, direction=Direction.ALL): - vertex_col, degree_col = graph_primtypes_wrapper._degree(self, direction) - df = cudf.DataFrame() - df["vertex"] = vertex_col - df["degree"] = degree_col - - if self.properties.renumbered: - # Get the internal vertex IDs - nodes = self.renumber_map.df_internal_to_external["id"] - else: - nodes = self.nodes() - # If the vertex IDs are not contiguous, remove results for the - # isolated vertices - df = df[df["vertex"].isin(nodes.to_cupy())] - - if vertex_subset is not None: - if not isinstance(vertex_subset, cudf.Series): - vertex_subset = cudf.Series(vertex_subset) - if self.properties.renumbered: - vertex_subset = self.renumber_map.to_internal_vertex_id( - vertex_subset - ) - vertex_subset = vertex_subset.to_cupy() - df = df[df["vertex"].isin(vertex_subset)] - - if self.properties.renumbered: - df = self.renumber_map.unrenumber(df, "vertex") - - return df + return self.degrees_function(vertex_subset, "degrees") def _make_plc_graph( self, diff --git a/python/cugraph/cugraph/tests/centrality/test_degree_centrality_mg.py b/python/cugraph/cugraph/tests/centrality/test_degree_centrality_mg.py index a46f4b9463b..1bef1e0872b 100644 --- a/python/cugraph/cugraph/tests/centrality/test_degree_centrality_mg.py +++ b/python/cugraph/cugraph/tests/centrality/test_degree_centrality_mg.py @@ -1,4 +1,4 @@ -# Copyright (c) 2018-2023, NVIDIA CORPORATION. +# Copyright (c) 2018-2024, 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 @@ -18,7 +18,6 @@ import cudf import dask_cudf import cugraph -from cugraph.dask.common.mg_utils import is_single_gpu from cugraph.testing.utils import RAPIDS_DATASET_ROOT_DIR_PATH from cudf.testing import assert_series_equal @@ -41,7 +40,6 @@ def setup_function(): @pytest.mark.mg -@pytest.mark.skipif(is_single_gpu(), reason="skipping MG testing on Single GPU system") @pytest.mark.parametrize("directed", IS_DIRECTED) @pytest.mark.parametrize("data_file", DATA_PATH) def test_dask_mg_degree(dask_client, directed, data_file): diff --git a/python/pylibcugraph/pylibcugraph/CMakeLists.txt b/python/pylibcugraph/pylibcugraph/CMakeLists.txt index c2e22fc1ff7..7cc90145949 100644 --- a/python/pylibcugraph/pylibcugraph/CMakeLists.txt +++ b/python/pylibcugraph/pylibcugraph/CMakeLists.txt @@ -1,5 +1,5 @@ # ============================================================================= -# Copyright (c) 2022-2023, NVIDIA CORPORATION. +# Copyright (c) 2022-2024, 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 @@ -57,6 +57,7 @@ set(cython_sources utils.pyx weakly_connected_components.pyx replicate_edgelist.pyx + degrees.pyx ) set(linked_libraries cugraph::cugraph;cugraph::cugraph_c) diff --git a/python/pylibcugraph/pylibcugraph/__init__.py b/python/pylibcugraph/pylibcugraph/__init__.py index ab518e24cae..dcdef05e106 100644 --- a/python/pylibcugraph/pylibcugraph/__init__.py +++ b/python/pylibcugraph/pylibcugraph/__init__.py @@ -95,6 +95,8 @@ from pylibcugraph.sorensen_coefficients import sorensen_coefficients +from pylibcugraph.degrees import in_degrees, out_degrees, degrees + from pylibcugraph import exceptions diff --git a/python/pylibcugraph/pylibcugraph/_cugraph_c/graph_functions.pxd b/python/pylibcugraph/pylibcugraph/_cugraph_c/graph_functions.pxd index 90bc041e5f0..6f1ac1f640b 100644 --- a/python/pylibcugraph/pylibcugraph/_cugraph_c/graph_functions.pxd +++ b/python/pylibcugraph/pylibcugraph/_cugraph_c/graph_functions.pxd @@ -182,3 +182,58 @@ cdef extern from "cugraph_c/graph_functions.h": cugraph_induced_subgraph_result_t** result, cugraph_error_t** error ) + + ########################################################################### + # degrees + ctypedef struct cugraph_degrees_result_t: + pass + + cdef cugraph_error_code_t \ + cugraph_in_degrees( + const cugraph_resource_handle_t* handle, + cugraph_graph_t* graph, + const cugraph_type_erased_device_array_view_t* source_vertices, + bool_t do_expensive_check, + cugraph_degrees_result_t** result, + cugraph_error_t** error + ) + + cdef cugraph_error_code_t \ + cugraph_out_degrees( + const cugraph_resource_handle_t* handle, + cugraph_graph_t* graph, + const cugraph_type_erased_device_array_view_t* source_vertices, + bool_t do_expensive_check, + cugraph_degrees_result_t** result, + cugraph_error_t** error + ) + + cdef cugraph_error_code_t \ + cugraph_degrees( + const cugraph_resource_handle_t* handle, + cugraph_graph_t* graph, + const cugraph_type_erased_device_array_view_t* source_vertices, + bool_t do_expensive_check, + cugraph_degrees_result_t** result, + cugraph_error_t** error + ) + + cdef cugraph_type_erased_device_array_view_t* \ + cugraph_degrees_result_get_vertices( + cugraph_degrees_result_t* degrees_result + ) + + cdef cugraph_type_erased_device_array_view_t* \ + cugraph_degrees_result_get_in_degrees( + cugraph_degrees_result_t* degrees_result + ) + + cdef cugraph_type_erased_device_array_view_t* \ + cugraph_degrees_result_get_out_degrees( + cugraph_degrees_result_t* degrees_result + ) + + cdef void \ + cugraph_degrees_result_free( + cugraph_degrees_result_t* degrees_result + ) diff --git a/python/pylibcugraph/pylibcugraph/degrees.pyx b/python/pylibcugraph/pylibcugraph/degrees.pyx new file mode 100644 index 00000000000..7818da441bd --- /dev/null +++ b/python/pylibcugraph/pylibcugraph/degrees.pyx @@ -0,0 +1,307 @@ +# Copyright (c) 2024, 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. + +# Have cython use python 3 syntax +# cython: language_level = 3 + +from libc.stdint cimport uintptr_t + +from pylibcugraph._cugraph_c.resource_handle cimport ( + bool_t, + data_type_id_t, + cugraph_resource_handle_t, +) +from pylibcugraph._cugraph_c.error cimport ( + cugraph_error_code_t, + cugraph_error_t, +) +from pylibcugraph._cugraph_c.array cimport ( + cugraph_type_erased_device_array_view_t, +) +from pylibcugraph._cugraph_c.graph cimport ( + cugraph_graph_t, +) +from pylibcugraph._cugraph_c.graph_functions cimport ( + cugraph_degrees_result_t, + cugraph_degrees, + cugraph_in_degrees, + cugraph_out_degrees, + cugraph_degrees_result_get_vertices, + cugraph_degrees_result_get_in_degrees, + cugraph_degrees_result_get_out_degrees, + cugraph_degrees_result_free, +) +from pylibcugraph.resource_handle cimport ( + ResourceHandle, +) +from pylibcugraph.graphs cimport ( + _GPUGraph, +) +from pylibcugraph.utils cimport ( + assert_success, + copy_to_cupy_array, + assert_CAI_type, + create_cugraph_type_erased_device_array_view_from_py_obj, +) + + +def in_degrees(ResourceHandle resource_handle, + _GPUGraph graph, + source_vertices, + bool_t do_expensive_check): + """ + Compute the in degrees for the nodes of the graph. + + Parameters + ---------- + resource_handle : ResourceHandle + Handle to the underlying device resources needed for referencing data + and running algorithms. + + graph : SGGraph or MGGraph + The input graph, for either Single or Multi-GPU operations. + + source_vertices : cupy array + The nodes for which we will compute degrees. + + do_expensive_check : bool_t + A flag to run expensive checks for input arguments if True. + + Returns + ------- + A tuple of device arrays, where the first item in the tuple is a device + array containing the vertices, the second item in the tuple is a device + array containing the in degrees for the vertices. + + Examples + -------- + >>> import pylibcugraph, cupy, numpy + >>> srcs = cupy.asarray([0, 1, 2], dtype=numpy.int32) + >>> dsts = cupy.asarray([1, 2, 3], dtype=numpy.int32) + >>> weights = cupy.asarray([1.0, 1.0, 1.0], dtype=numpy.float32) + >>> resource_handle = pylibcugraph.ResourceHandle() + >>> graph_props = pylibcugraph.GraphProperties( + ... is_symmetric=False, is_multigraph=False) + >>> G = pylibcugraph.SGGraph( + ... resource_handle, graph_props, srcs, dsts, weight_array=weights, + ... store_transposed=True, renumber=False, do_expensive_check=False) + >>> (vertices, in_degrees) = pylibcugraph.in_degrees( + resource_handle, G, None, False) + + """ + + cdef cugraph_resource_handle_t* c_resource_handle_ptr = \ + resource_handle.c_resource_handle_ptr + cdef cugraph_graph_t* c_graph_ptr = graph.c_graph_ptr + + cdef cugraph_degrees_result_t* result_ptr + cdef cugraph_error_code_t error_code + cdef cugraph_error_t* error_ptr + + assert_CAI_type(source_vertices, "source_vertices", True) + + cdef cugraph_type_erased_device_array_view_t* \ + source_vertices_ptr = \ + create_cugraph_type_erased_device_array_view_from_py_obj( + source_vertices) + + error_code = cugraph_in_degrees(c_resource_handle_ptr, + c_graph_ptr, + source_vertices_ptr, + do_expensive_check, + &result_ptr, + &error_ptr) + assert_success(error_code, error_ptr, "cugraph_in_degrees") + + # Extract individual device array pointers from result and copy to cupy + # arrays for returning. + cdef cugraph_type_erased_device_array_view_t* vertices_ptr = \ + cugraph_degrees_result_get_vertices(result_ptr) + cdef cugraph_type_erased_device_array_view_t* in_degrees_ptr = \ + cugraph_degrees_result_get_in_degrees(result_ptr) + + cupy_vertices = copy_to_cupy_array(c_resource_handle_ptr, vertices_ptr) + cupy_in_degrees = copy_to_cupy_array(c_resource_handle_ptr, in_degrees_ptr) + + cugraph_degrees_result_free(result_ptr) + + return (cupy_vertices, cupy_in_degrees) + +def out_degrees(ResourceHandle resource_handle, + _GPUGraph graph, + source_vertices, + bool_t do_expensive_check): + """ + Compute the out degrees for the nodes of the graph. + + Parameters + ---------- + resource_handle : ResourceHandle + Handle to the underlying device resources needed for referencing data + and running algorithms. + + graph : SGGraph or MGGraph + The input graph, for either Single or Multi-GPU operations. + + source_vertices : cupy array + The nodes for which we will compute degrees. + + do_expensive_check : bool_t + A flag to run expensive checks for input arguments if True. + + Returns + ------- + A tuple of device arrays, where the first item in the tuple is a device + array containing the vertices, the second item in the tuple is a device + array containing the out degrees for the vertices. + + Examples + -------- + >>> import pylibcugraph, cupy, numpy + >>> srcs = cupy.asarray([0, 1, 2], dtype=numpy.int32) + >>> dsts = cupy.asarray([1, 2, 3], dtype=numpy.int32) + >>> weights = cupy.asarray([1.0, 1.0, 1.0], dtype=numpy.float32) + >>> resource_handle = pylibcugraph.ResourceHandle() + >>> graph_props = pylibcugraph.GraphProperties( + ... is_symmetric=False, is_multigraph=False) + >>> G = pylibcugraph.SGGraph( + ... resource_handle, graph_props, srcs, dsts, weight_array=weights, + ... store_transposed=True, renumber=False, do_expensive_check=False) + >>> (vertices, out_degrees) = pylibcugraph.out_degrees( + resource_handle, G, None, False) + + """ + + cdef cugraph_resource_handle_t* c_resource_handle_ptr = \ + resource_handle.c_resource_handle_ptr + cdef cugraph_graph_t* c_graph_ptr = graph.c_graph_ptr + + cdef cugraph_degrees_result_t* result_ptr + cdef cugraph_error_code_t error_code + cdef cugraph_error_t* error_ptr + + assert_CAI_type(source_vertices, "source_vertices", True) + + cdef cugraph_type_erased_device_array_view_t* \ + source_vertices_ptr = \ + create_cugraph_type_erased_device_array_view_from_py_obj( + source_vertices) + + error_code = cugraph_out_degrees(c_resource_handle_ptr, + c_graph_ptr, + source_vertices_ptr, + do_expensive_check, + &result_ptr, + &error_ptr) + assert_success(error_code, error_ptr, "cugraph_out_degrees") + + # Extract individual device array pointers from result and copy to cupy + # arrays for returning. + cdef cugraph_type_erased_device_array_view_t* vertices_ptr = \ + cugraph_degrees_result_get_vertices(result_ptr) + cdef cugraph_type_erased_device_array_view_t* out_degrees_ptr = \ + cugraph_degrees_result_get_out_degrees(result_ptr) + + cupy_vertices = copy_to_cupy_array(c_resource_handle_ptr, vertices_ptr) + cupy_out_degrees = copy_to_cupy_array(c_resource_handle_ptr, out_degrees_ptr) + + cugraph_degrees_result_free(result_ptr) + + return (cupy_vertices, cupy_out_degrees) + + +def degrees(ResourceHandle resource_handle, + _GPUGraph graph, + source_vertices, + bool_t do_expensive_check): + """ + Compute the degrees for the nodes of the graph. + + Parameters + ---------- + resource_handle : ResourceHandle + Handle to the underlying device resources needed for referencing data + and running algorithms. + + graph : SGGraph or MGGraph + The input graph, for either Single or Multi-GPU operations. + + source_vertices : cupy array + The nodes for which we will compute degrees. + + do_expensive_check : bool_t + A flag to run expensive checks for input arguments if True. + + Returns + ------- + A tuple of device arrays, where the first item in the tuple is a device + array containing the vertices, the second item in the tuple is a device + array containing the in degrees for the vertices, the third item in the + tuple is a device array containing the out degrees for the vertices. + + Examples + -------- + >>> import pylibcugraph, cupy, numpy + >>> srcs = cupy.asarray([0, 1, 2], dtype=numpy.int32) + >>> dsts = cupy.asarray([1, 2, 3], dtype=numpy.int32) + >>> weights = cupy.asarray([1.0, 1.0, 1.0], dtype=numpy.float32) + >>> resource_handle = pylibcugraph.ResourceHandle() + >>> graph_props = pylibcugraph.GraphProperties( + ... is_symmetric=False, is_multigraph=False) + >>> G = pylibcugraph.SGGraph( + ... resource_handle, graph_props, srcs, dsts, weight_array=weights, + ... store_transposed=True, renumber=False, do_expensive_check=False) + >>> (vertices, in_degrees, out_degrees) = pylibcugraph.degrees( + resource_handle, G, None, False) + + """ + + cdef cugraph_resource_handle_t* c_resource_handle_ptr = \ + resource_handle.c_resource_handle_ptr + cdef cugraph_graph_t* c_graph_ptr = graph.c_graph_ptr + + cdef cugraph_degrees_result_t* result_ptr + cdef cugraph_error_code_t error_code + cdef cugraph_error_t* error_ptr + + assert_CAI_type(source_vertices, "source_vertices", True) + + cdef cugraph_type_erased_device_array_view_t* \ + source_vertices_ptr = \ + create_cugraph_type_erased_device_array_view_from_py_obj( + source_vertices) + + error_code = cugraph_degrees(c_resource_handle_ptr, + c_graph_ptr, + source_vertices_ptr, + do_expensive_check, + &result_ptr, + &error_ptr) + assert_success(error_code, error_ptr, "cugraph_degrees") + + # Extract individual device array pointers from result and copy to cupy + # arrays for returning. + cdef cugraph_type_erased_device_array_view_t* vertices_ptr = \ + cugraph_degrees_result_get_vertices(result_ptr) + cdef cugraph_type_erased_device_array_view_t* in_degrees_ptr = \ + cugraph_degrees_result_get_in_degrees(result_ptr) + cdef cugraph_type_erased_device_array_view_t* out_degrees_ptr = \ + cugraph_degrees_result_get_out_degrees(result_ptr) + + cupy_vertices = copy_to_cupy_array(c_resource_handle_ptr, vertices_ptr) + cupy_in_degrees = copy_to_cupy_array(c_resource_handle_ptr, in_degrees_ptr) + cupy_out_degrees = copy_to_cupy_array(c_resource_handle_ptr, out_degrees_ptr) + + cugraph_degrees_result_free(result_ptr) + + return (cupy_vertices, cupy_in_degrees, cupy_out_degrees)