Skip to content

Replace rmm::device_vector & thrust::host_vector with rmm::device_uvector & std::vector, respectively.#1421

Merged
rapids-bot[bot] merged 4 commits intorapidsai:branch-0.19from
seunghwak:enh_uvector
Mar 4, 2021
Merged

Replace rmm::device_vector & thrust::host_vector with rmm::device_uvector & std::vector, respectively.#1421
rapids-bot[bot] merged 4 commits intorapidsai:branch-0.19from
seunghwak:enh_uvector

Conversation

@seunghwak
Copy link
Contributor

@seunghwak seunghwak commented Feb 22, 2021

  • Replace rmm::device_vector with rmm::device_uvector for better concurrency in multi-stream executions
  • Replace thrust::host_vector with std::vector

This PR partially addresses #1390

@seunghwak seunghwak requested a review from a team as a code owner February 22, 2021 18:57
@seunghwak
Copy link
Contributor Author

@afender you may want to review this.

@seunghwak seunghwak added 3 - Ready for Review improvement Improvement / enhancement to an existing function non-breaking Non-breaking change labels Feb 22, 2021
Copy link
Member

@afender afender left a comment

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Looks good but :

  • consider using the latest streams features from RAFT
  • could we think of a test to verify it is now executing on the stream?

bucket_sizes_device_ptr,
VertexFrontierType::kNumBuckets,
handle.get_stream());
CUDA_TRY(cudaStreamSynchronize(handle.get_stream()));
Copy link
Member

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Do we really need this sync?

If so, consider handle.get_stream_view.synchronize() instead.
Or handle.wait_on_user_stream but this triggers worker stream sync as well

Copy link
Contributor Author

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

This is necessary as without this, bucket_sizes (which is std::vector and is not aware of stream) is not guaranteed to hold up-to-date data and is used right after this sync.

CUDA_TRY(cudaStreamSynchronize(handle.get_stream())); this is also used all around the codebase, so I will create an issue to replace all at once.

: handle_ptr_(&handle), elements_(capacity, invalid_vertex_id<vertex_t>::value)
: handle_ptr_(&handle), elements_(capacity, handle.get_stream())
{
thrust::fill(rmm::exec_policy(handle_ptr_->get_stream())->on(handle_ptr_->get_stream()),
Copy link
Member

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Consider replacing
thrust::fill(rmm::exec_policy(handle_ptr_->get_stream())->on(handle_ptr_->get_stream()),
by
thrust::fill(rmm::exec_policy(handle.get_stream_view()),

This applies to all thrust calls

Copy link
Contributor Author

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Yes, same story here. This should be replaced all at once.

@seunghwak
Copy link
Contributor Author

@afender #1422 is created to transition cuGraph to rmm's new cuda_stream_view.

@seunghwak
Copy link
Contributor Author

  • could we think of a test to verify it is now executing on the stream?

This is tricky I guess; the first thing is this is not a black-or-white thing and this will affect only performance not correctness.

I guess we need cuGraph-wise benchmark scheme to track multi-stream execution of (every) graph analytics?

@codecov-io
Copy link

codecov-io commented Feb 22, 2021

Codecov Report

❗ No coverage uploaded for pull request base (branch-0.19@3f13ffc). Click here to learn what that means.
The diff coverage is n/a.

Impacted file tree graph

@@              Coverage Diff               @@
##             branch-0.19    #1421   +/-   ##
==============================================
  Coverage               ?   60.75%           
==============================================
  Files                  ?       70           
  Lines                  ?     3134           
  Branches               ?        0           
==============================================
  Hits                   ?     1904           
  Misses                 ?     1230           
  Partials               ?        0           

Continue to review full report at Codecov.

Legend - Click here to learn more
Δ = absolute <relative> (impact), ø = not affected, ? = missing data
Powered by Codecov. Last update 3f13ffc...9350f96. Read the comment docs.

@afender
Copy link
Member

afender commented Feb 26, 2021

  • could we think of a test to verify it is now executing on the stream?

This is tricky I guess; the first thing is this is not a black-or-white thing and this will affect only performance not correctness.

I guess we need cuGraph-wise benchmark scheme to track multi-stream execution of (every) graph analytics?

Ok, for now I can rebase my concurrency branch and generate a profile after this gets merged.

@BradReesWork BradReesWork requested a review from aschaffer March 1, 2021 19:03
@BradReesWork BradReesWork added this to the 0.19 milestone Mar 1, 2021
@BradReesWork
Copy link
Member

@gpucibot merge

@rapids-bot rapids-bot bot merged commit c1047ed into rapidsai:branch-0.19 Mar 4, 2021
@seunghwak seunghwak deleted the enh_uvector branch June 24, 2021 19:03
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment

Labels

improvement Improvement / enhancement to an existing function non-breaking Non-breaking change

Projects

None yet

Development

Successfully merging this pull request may close these issues.

4 participants