Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Jump to bottom

Separate edge weights from graph objects and update primitives to support general edge properties. #2843

Merged
merged 41 commits into from
Nov 21, 2022
Merged

Separate edge weights from graph objects and update primitives to support general edge properties. #2843

Show file tree
Hide file tree
Changes from 34 commits
Commits
Show all changes
41 commits
Select commit Hold shift + click to select a range
2819834
delete dead code
seunghwak Oct 20, 2022
8dde153
code cleanup
seunghwak Oct 20, 2022
83b1707
remove SG graph_test
seunghwak Oct 20, 2022
3262c1f
delete dead comment
seunghwak Oct 21, 2022
4bf057e
Merge branch 'branch-22.12' of github.com:rapidsai/cugraph into enh_g…
seunghwak Oct 21, 2022
a39d9fa
undo deleting doxygen hints, delete the serializer_t friend declaration
seunghwak Oct 21, 2022
fcbe57a
delete unnecessary space
seunghwak Oct 21, 2022
7bdae42
Merge branch 'branch-22.12' of github.com:rapidsai/cugraph into enh_g…
seunghwak Oct 21, 2022
b381ae9
move graph_t member functions related to weight_t to graph_functions.hpp
seunghwak Oct 22, 2022
13e82b2
Merge branch 'branch-22.12' of github.com:rapidsai/cugraph into enh_m…
seunghwak Oct 22, 2022
8ab47fb
delete dead code
seunghwak Oct 22, 2022
ab3972d
delete unnecessary includes
seunghwak Oct 22, 2022
6d9ebcf
move graph_view_t member functions related to weight_t to graph_funct…
seunghwak Oct 22, 2022
be64ac3
resolve merge conflicts
seunghwak Oct 24, 2022
3f97489
clang-format
seunghwak Oct 24, 2022
f0a7b9e
Merge branch 'upstream_pr2841' into fea_prim_edge_property
seunghwak Oct 24, 2022
d1f4835
separate weights from graph objects in graph creation
seunghwak Oct 24, 2022
11ddd14
update graph_t & graph_view_t and per_v_transform_reduce_incoming|out…
seunghwak Oct 24, 2022
ddc8d79
update bfs & transform_reduce_v_frontier_outgoing_e
seunghwak Oct 24, 2022
a9bfafe
update graph_functions.hpp
seunghwak Oct 25, 2022
1a5810d
update algorithms.hpp
seunghwak Oct 25, 2022
382bc90
update more algorithms and primitves to work with separate graph & ed…
seunghwak Oct 25, 2022
684a7c4
update hits, random walks, and random sampling primitive
seunghwak Oct 26, 2022
8e4c0cc
resolve merge conflicts
seunghwak Oct 26, 2022
2523aad
update more algorithms and primitives
seunghwak Oct 28, 2022
5d86cdb
Merge branch 'branch-22.12' of github.com:rapidsai/cugraph into fea_p…
seunghwak Oct 28, 2022
02ef726
update smililarity measure algorithms
seunghwak Oct 28, 2022
068deba
fix compile errors in building libcugraph.so (excluding cython.cu)
seunghwak Oct 30, 2022
a0c7d33
additional updates in the core C++ part
seunghwak Oct 31, 2022
e91d8ed
update the C++ side of the C API
seunghwak Oct 31, 2022
604de86
update MG C++ tests
seunghwak Oct 31, 2022
4decd58
fix undefined symbol error
seunghwak Oct 31, 2022
9506555
bug fix
seunghwak Oct 31, 2022
0306d25
additional bug fix
seunghwak Nov 1, 2022
e598d44
resolve merge conflicts
seunghwak Nov 18, 2022
7078e16
fix compile errors
seunghwak Nov 18, 2022
6b53db4
clang-format & copyright year fixes
seunghwak Nov 18, 2022
1537d4d
bug fix in MG random walk test code
seunghwak Nov 18, 2022
e4aa3b4
delete unused files
seunghwak Nov 18, 2022
0b27e9a
bug fix in SG random walk test code
seunghwak Nov 18, 2022
4ed9e1a
Merge branch 'branch-22.12' of github.com:rapidsai/cugraph into fea_p…
seunghwak Nov 18, 2022
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
269 changes: 151 additions & 118 deletions cpp/include/cugraph/algorithms.hpp
View file
Open in desktop

Large diffs are not rendered by default.

358 changes: 19 additions & 339 deletions cpp/include/cugraph/detail/decompress_edge_partition.cuh
View file
Open in desktop

Large diffs are not rendered by default.

57 changes: 16 additions & 41 deletions cpp/include/cugraph/edge_partition_device_view.cuh
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -92,16 +92,12 @@ struct local_degree_op_t {
}
};

template <typename vertex_t, typename edge_t, typename weight_t>
template <typename vertex_t, typename edge_t>
class edge_partition_device_view_base_t {
public:
edge_partition_device_view_base_t(raft::device_span<edge_t const> offsets,
raft::device_span<vertex_t const> indices,
std::optional<raft::device_span<weight_t const>> weights)
: offsets_(offsets),
indices_(indices),
weights_(weights ? thrust::optional<raft::device_span<weight_t const>>(*weights)
: thrust::nullopt)
raft::device_span<vertex_t const> indices)
: offsets_(offsets), indices_(indices)
{
}

Expand All @@ -112,21 +108,15 @@ class edge_partition_device_view_base_t {

__host__ __device__ edge_t const* offsets() const { return offsets_.data(); }
__host__ __device__ vertex_t const* indices() const { return indices_.data(); }
__host__ __device__ thrust::optional<weight_t const*> weights() const
{
return weights_ ? thrust::optional<weight_t const*>{(*weights_).data()} : thrust::nullopt;
}

// major_idx == major offset if CSR/CSC, major_offset != major_idx if DCSR/DCSC
__device__ thrust::tuple<vertex_t const*, thrust::optional<weight_t const*>, edge_t> local_edges(
__device__ thrust::tuple<vertex_t const*, edge_t, edge_t> local_edges(
vertex_t major_idx) const noexcept
{
auto edge_offset = offsets_[major_idx];
auto local_degree = offsets_[major_idx + 1] - edge_offset;
auto indices = indices_.data() + edge_offset;
auto weights = weights_ ? thrust::optional<weight_t const*>{(*weights_).data() + edge_offset}
: thrust::nullopt;
return thrust::make_tuple(indices, weights, local_degree);
return thrust::make_tuple(indices, edge_offset, local_degree);
}

// major_idx == major offset if CSR/CSC, major_offset != major_idx if DCSR/DCSC
Expand All @@ -142,30 +132,20 @@ class edge_partition_device_view_base_t {
// should be trivially copyable to device
raft::device_span<edge_t const> offsets_{nullptr};
raft::device_span<vertex_t const> indices_{nullptr};
thrust::optional<raft::device_span<weight_t const>> weights_{thrust::nullopt};
};

} // namespace detail

template <typename vertex_t,
typename edge_t,
typename weight_t,
bool multi_gpu,
typename Enable = void>
template <typename vertex_t, typename edge_t, bool multi_gpu, typename Enable = void>
class edge_partition_device_view_t;

// multi-GPU version
template <typename vertex_t, typename edge_t, typename weight_t, bool multi_gpu>
class edge_partition_device_view_t<vertex_t,
edge_t,
weight_t,
multi_gpu,
std::enable_if_t<multi_gpu>>
: public detail::edge_partition_device_view_base_t<vertex_t, edge_t, weight_t> {
template <typename vertex_t, typename edge_t, bool multi_gpu>
class edge_partition_device_view_t<vertex_t, edge_t, multi_gpu, std::enable_if_t<multi_gpu>>
: public detail::edge_partition_device_view_base_t<vertex_t, edge_t> {
public:
edge_partition_device_view_t(edge_partition_view_t<vertex_t, edge_t, weight_t, multi_gpu> view)
: detail::edge_partition_device_view_base_t<vertex_t, edge_t, weight_t>(
view.offsets(), view.indices(), view.weights()),
edge_partition_device_view_t(edge_partition_view_t<vertex_t, edge_t, multi_gpu> view)
: detail::edge_partition_device_view_base_t<vertex_t, edge_t>(view.offsets(), view.indices()),
dcs_nzd_vertices_(detail::to_thrust_optional(view.dcs_nzd_vertices())),
major_hypersparse_first_(detail::to_thrust_optional(view.major_hypersparse_first())),
major_range_first_(view.major_range_first()),
Expand Down Expand Up @@ -359,17 +339,12 @@ class edge_partition_device_view_t<vertex_t,
};

// single-GPU version
template <typename vertex_t, typename edge_t, typename weight_t, bool multi_gpu>
class edge_partition_device_view_t<vertex_t,
edge_t,
weight_t,
multi_gpu,
std::enable_if_t<!multi_gpu>>
: public detail::edge_partition_device_view_base_t<vertex_t, edge_t, weight_t> {
template <typename vertex_t, typename edge_t, bool multi_gpu>
class edge_partition_device_view_t<vertex_t, edge_t, multi_gpu, std::enable_if_t<!multi_gpu>>
: public detail::edge_partition_device_view_base_t<vertex_t, edge_t> {
public:
edge_partition_device_view_t(edge_partition_view_t<vertex_t, edge_t, weight_t, multi_gpu> view)
: detail::edge_partition_device_view_base_t<vertex_t, edge_t, weight_t>(
view.offsets(), view.indices(), view.weights()),
edge_partition_device_view_t(edge_partition_view_t<vertex_t, edge_t, multi_gpu> view)
: detail::edge_partition_device_view_base_t<vertex_t, edge_t>(view.offsets(), view.indices()),
number_of_vertices_(view.major_range_last())
{
}
Expand Down
2 changes: 2 additions & 0 deletions cpp/include/cugraph/edge_partition_edge_property_device_view.cuh
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -39,6 +39,8 @@ class edge_partition_edge_property_device_view_t {
value_first_ = view.value_firsts()[partition_idx];
}

__host__ __device__ ValueIterator value_first() { return value_first_; }

__device__ ValueIterator get_iter(edge_t offset) const { return value_first_ + offset; }

__device__ value_type get(edge_t offset) const { return *get_iter(offset); }
Expand Down
33 changes: 12 additions & 21 deletions cpp/include/cugraph/edge_partition_view.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -24,53 +24,45 @@ namespace cugraph {

namespace detail {

template <typename vertex_t, typename edge_t, typename weight_t>
template <typename vertex_t, typename edge_t>
class edge_partition_view_base_t {
public:
edge_partition_view_base_t(raft::device_span<edge_t const> offsets,
raft::device_span<vertex_t const> indices,
std::optional<raft::device_span<weight_t const>> weights)
: offsets_(offsets), indices_(indices), weights_(weights)
raft::device_span<vertex_t const> indices)
: offsets_(offsets), indices_(indices)
{
}

edge_t number_of_edges() const { return static_cast<edge_t>(indices_.size()); }

raft::device_span<edge_t const> offsets() const { return offsets_; }
raft::device_span<vertex_t const> indices() const { return indices_; }
std::optional<raft::device_span<weight_t const>> weights() const { return weights_; }

private:
raft::device_span<edge_t const> offsets_{};
raft::device_span<vertex_t const> indices_{};
std::optional<raft::device_span<weight_t const>> weights_{std::nullopt};
};

} // namespace detail

template <typename vertex_t,
typename edge_t,
typename weight_t,
bool multi_gpu,
typename Enable = void>
template <typename vertex_t, typename edge_t, bool multi_gpu, typename Enable = void>
class edge_partition_view_t;

// multi-GPU version
template <typename vertex_t, typename edge_t, typename weight_t, bool multi_gpu>
class edge_partition_view_t<vertex_t, edge_t, weight_t, multi_gpu, std::enable_if_t<multi_gpu>>
: public detail::edge_partition_view_base_t<vertex_t, edge_t, weight_t> {
template <typename vertex_t, typename edge_t, bool multi_gpu>
class edge_partition_view_t<vertex_t, edge_t, multi_gpu, std::enable_if_t<multi_gpu>>
: public detail::edge_partition_view_base_t<vertex_t, edge_t> {
public:
edge_partition_view_t(raft::device_span<edge_t const> offsets,
raft::device_span<vertex_t const> indices,
std::optional<raft::device_span<weight_t const>> weights,
std::optional<raft::device_span<vertex_t const>> dcs_nzd_vertices,
std::optional<vertex_t> major_hypersparse_first,
vertex_t major_range_first,
vertex_t major_range_last,
vertex_t minor_range_first,
vertex_t minor_range_last,
vertex_t major_value_start_offset)
: detail::edge_partition_view_base_t<vertex_t, edge_t, weight_t>(offsets, indices, weights),
: detail::edge_partition_view_base_t<vertex_t, edge_t>(offsets, indices),
dcs_nzd_vertices_(dcs_nzd_vertices),
major_hypersparse_first_(major_hypersparse_first),
major_range_first_(major_range_first),
Expand Down Expand Up @@ -109,15 +101,14 @@ class edge_partition_view_t<vertex_t, edge_t, weight_t, multi_gpu, std::enable_i
};

// single-GPU version
template <typename vertex_t, typename edge_t, typename weight_t, bool multi_gpu>
class edge_partition_view_t<vertex_t, edge_t, weight_t, multi_gpu, std::enable_if_t<!multi_gpu>>
: public detail::edge_partition_view_base_t<vertex_t, edge_t, weight_t> {
template <typename vertex_t, typename edge_t, bool multi_gpu>
class edge_partition_view_t<vertex_t, edge_t, multi_gpu, std::enable_if_t<!multi_gpu>>
: public detail::edge_partition_view_base_t<vertex_t, edge_t> {
public:
edge_partition_view_t(raft::device_span<edge_t const> offsets,
raft::device_span<vertex_t const> indices,
std::optional<raft::device_span<weight_t const>> weights,
vertex_t number_of_vertices)
: detail::edge_partition_view_base_t<vertex_t, edge_t, weight_t>(offsets, indices, weights),
: detail::edge_partition_view_base_t<vertex_t, edge_t>(offsets, indices),
number_of_vertices_(number_of_vertices)
{
}
Expand Down
14 changes: 5 additions & 9 deletions cpp/include/cugraph/edge_property.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -27,8 +27,6 @@

namespace cugraph {

namespace detail {

template <typename edge_t, typename ValueIterator>
class edge_property_view_t {
public:
Expand Down Expand Up @@ -59,8 +57,6 @@ class edge_dummy_property_view_t {
using value_iterator = void*;
};

} // namespace detail

template <typename GraphViewType, typename T>
class edge_property_t {
public:
Expand Down Expand Up @@ -98,8 +94,8 @@ class edge_property_t {
edge_partition_edge_counts[i] = size_dataframe_buffer(buffers_[i]);
}

return detail::edge_property_view_t<edge_type, const_value_iterator>(
edge_partition_value_firsts, edge_partition_edge_counts);
return edge_property_view_t<edge_type, const_value_iterator>(edge_partition_value_firsts,
edge_partition_edge_counts);
}

auto mutable_view()
Expand All @@ -113,8 +109,8 @@ class edge_property_t {
edge_partition_edge_counts[i] = size_dataframe_buffer(buffers_[i]);
}

return detail::edge_property_view_t<edge_type, value_iterator>(edge_partition_value_firsts,
edge_partition_edge_counts);
return edge_property_view_t<edge_type, value_iterator>(edge_partition_value_firsts,
edge_partition_edge_counts);
}

private:
Expand All @@ -125,7 +121,7 @@ class edge_dummy_property_t {
public:
using value_type = thrust::nullopt_t;

auto view() const { return detail::edge_dummy_property_view_t{}; }
auto view() const { return edge_dummy_property_view_t{}; }
};

} // namespace cugraph
96 changes: 32 additions & 64 deletions cpp/include/cugraph/eidecl_graph.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -16,68 +16,36 @@
#pragma once

namespace cugraph {
extern template class graph_t<int32_t, int32_t, float, true, true, void>;
extern template class graph_t<int32_t, int32_t, float, true, false, void>;
extern template class graph_t<int32_t, int32_t, float, false, true, void>;
extern template class graph_t<int32_t, int32_t, float, false, false, void>;
extern template class graph_t<int32_t, int32_t, double, true, true, void>;
extern template class graph_t<int32_t, int32_t, double, true, false, void>;
extern template class graph_t<int32_t, int32_t, double, false, true, void>;
extern template class graph_t<int32_t, int32_t, double, false, false, void>;
extern template class graph_t<int32_t, int64_t, float, true, true, void>;
extern template class graph_t<int32_t, int64_t, float, true, false, void>;
extern template class graph_t<int32_t, int64_t, float, false, true, void>;
extern template class graph_t<int32_t, int64_t, float, false, false, void>;
extern template class graph_t<int32_t, int64_t, double, true, true, void>;
extern template class graph_t<int32_t, int64_t, double, true, false, void>;
extern template class graph_t<int32_t, int64_t, double, false, true, void>;
extern template class graph_t<int32_t, int64_t, double, false, false, void>;
extern template class graph_t<int64_t, int32_t, float, true, true, void>;
extern template class graph_t<int64_t, int32_t, float, true, false, void>;
extern template class graph_t<int64_t, int32_t, float, false, true, void>;
extern template class graph_t<int64_t, int32_t, float, false, false, void>;
extern template class graph_t<int64_t, int32_t, double, true, true, void>;
extern template class graph_t<int64_t, int32_t, double, true, false, void>;
extern template class graph_t<int64_t, int32_t, double, false, true, void>;
extern template class graph_t<int64_t, int32_t, double, false, false, void>;
extern template class graph_t<int64_t, int64_t, float, true, true, void>;
extern template class graph_t<int64_t, int64_t, float, true, false, void>;
extern template class graph_t<int64_t, int64_t, float, false, true, void>;
extern template class graph_t<int64_t, int64_t, float, false, false, void>;
extern template class graph_t<int64_t, int64_t, double, true, true, void>;
extern template class graph_t<int64_t, int64_t, double, true, false, void>;
extern template class graph_t<int64_t, int64_t, double, false, true, void>;
extern template class graph_t<int64_t, int64_t, double, false, false, void>;
extern template class graph_view_t<int32_t, int32_t, float, true, true, void>;
extern template class graph_view_t<int32_t, int32_t, float, true, false, void>;
extern template class graph_view_t<int32_t, int32_t, float, false, true, void>;
extern template class graph_view_t<int32_t, int32_t, float, false, false, void>;
extern template class graph_view_t<int32_t, int32_t, double, true, true, void>;
extern template class graph_view_t<int32_t, int32_t, double, true, false, void>;
extern template class graph_view_t<int32_t, int32_t, double, false, true, void>;
extern template class graph_view_t<int32_t, int32_t, double, false, false, void>;
extern template class graph_view_t<int32_t, int64_t, float, true, true, void>;
extern template class graph_view_t<int32_t, int64_t, float, true, false, void>;
extern template class graph_view_t<int32_t, int64_t, float, false, true, void>;
extern template class graph_view_t<int32_t, int64_t, float, false, false, void>;
extern template class graph_view_t<int32_t, int64_t, double, true, true, void>;
extern template class graph_view_t<int32_t, int64_t, double, true, false, void>;
extern template class graph_view_t<int32_t, int64_t, double, false, true, void>;
extern template class graph_view_t<int32_t, int64_t, double, false, false, void>;
extern template class graph_view_t<int64_t, int32_t, float, true, true, void>;
extern template class graph_view_t<int64_t, int32_t, float, true, false, void>;
extern template class graph_view_t<int64_t, int32_t, float, false, true, void>;
extern template class graph_view_t<int64_t, int32_t, float, false, false, void>;
extern template class graph_view_t<int64_t, int32_t, double, true, true, void>;
extern template class graph_view_t<int64_t, int32_t, double, true, false, void>;
extern template class graph_view_t<int64_t, int32_t, double, false, true, void>;
extern template class graph_view_t<int64_t, int32_t, double, false, false, void>;
extern template class graph_view_t<int64_t, int64_t, float, true, true, void>;
extern template class graph_view_t<int64_t, int64_t, float, true, false, void>;
extern template class graph_view_t<int64_t, int64_t, float, false, true, void>;
extern template class graph_view_t<int64_t, int64_t, float, false, false, void>;
extern template class graph_view_t<int64_t, int64_t, double, true, true, void>;
extern template class graph_view_t<int64_t, int64_t, double, true, false, void>;
extern template class graph_view_t<int64_t, int64_t, double, false, true, void>;
extern template class graph_view_t<int64_t, int64_t, double, false, false, void>;
extern template class graph_t<int32_t, int32_t, true, true, void>;
extern template class graph_t<int32_t, int32_t, true, false, void>;
extern template class graph_t<int32_t, int32_t, false, true, void>;
extern template class graph_t<int32_t, int32_t, false, false, void>;
extern template class graph_t<int32_t, int64_t, true, true, void>;
extern template class graph_t<int32_t, int64_t, true, false, void>;
extern template class graph_t<int32_t, int64_t, false, true, void>;
extern template class graph_t<int32_t, int64_t, false, false, void>;
extern template class graph_t<int64_t, int32_t, true, true, void>;
extern template class graph_t<int64_t, int32_t, true, false, void>;
extern template class graph_t<int64_t, int32_t, false, true, void>;
extern template class graph_t<int64_t, int32_t, false, false, void>;
extern template class graph_t<int64_t, int64_t, true, true, void>;
extern template class graph_t<int64_t, int64_t, true, false, void>;
extern template class graph_t<int64_t, int64_t, false, true, void>;
extern template class graph_t<int64_t, int64_t, false, false, void>;
extern template class graph_view_t<int32_t, int32_t, true, true, void>;
extern template class graph_view_t<int32_t, int32_t, true, false, void>;
extern template class graph_view_t<int32_t, int32_t, false, true, void>;
extern template class graph_view_t<int32_t, int32_t, false, false, void>;
extern template class graph_view_t<int32_t, int64_t, true, true, void>;
extern template class graph_view_t<int32_t, int64_t, true, false, void>;
extern template class graph_view_t<int32_t, int64_t, false, true, void>;
extern template class graph_view_t<int32_t, int64_t, false, false, void>;
extern template class graph_view_t<int64_t, int32_t, true, true, void>;
extern template class graph_view_t<int64_t, int32_t, true, false, void>;
extern template class graph_view_t<int64_t, int32_t, false, true, void>;
extern template class graph_view_t<int64_t, int32_t, false, false, void>;
extern template class graph_view_t<int64_t, int64_t, true, true, void>;
extern template class graph_view_t<int64_t, int64_t, true, false, void>;
extern template class graph_view_t<int64_t, int64_t, false, true, void>;
extern template class graph_view_t<int64_t, int64_t, false, false, void>;
} // namespace cugraph
Loading