Skip to content

[SYCL][CUDA][PI] Add CUDA-specific pi_mem_advice values #5090

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

Merged
merged 7 commits into from
Feb 9, 2022
Merged

[SYCL][CUDA][PI] Add CUDA-specific pi_mem_advice values #5090

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
7 commits
Select commit Hold shift + click to select a range
df30b80
[SYCL][CUDA][PI] Added CUDA-specific pi_mem_advice values
t4c1 Dec 6, 2021
2a8b13d
format
t4c1 Dec 6, 2021
29a3503
addressed review comments
t4c1 Dec 6, 2021
34fba4a
format
t4c1 Dec 6, 2021
856e634
renamed advices to start with PI_MEM_ADVICE
t4c1 Feb 3, 2022
d85c8db
Merge branch 'cuda_memadvise' of https://github.com/t4c1/llvm into cu…
t4c1 Feb 3, 2022
adcedc5
Merge commit '8aa15c14dd0dc80dabe7f580ed191613dc6f33cf' into cuda_mem…
t4c1 Feb 4, 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
4 changes: 2 additions & 2 deletions sycl/include/CL/sycl/detail/cg.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -444,11 +444,11 @@ class CGAdviseUSM : public CG {
pi_mem_advice getAdvice() {
auto ExtendedMembers = getExtendedMembers();
if (!ExtendedMembers)
return PI_MEM_ADVISE_UNKNOWN;
return PI_MEM_ADVICE_UNKNOWN;
for (const ExtendedMemberT &EM : *ExtendedMembers)
if ((ExtendedMembersType::HANDLER_MEM_ADVICE == EM.MType) && EM.MData)
return *std::static_pointer_cast<pi_mem_advice>(EM.MData);
return PI_MEM_ADVISE_UNKNOWN;
return PI_MEM_ADVICE_UNKNOWN;
}
};

Expand Down
12 changes: 11 additions & 1 deletion sycl/include/CL/sycl/detail/pi.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -445,7 +445,17 @@ typedef enum {

typedef enum {
// Device-specific value opaque in PI API.
PI_MEM_ADVISE_UNKNOWN
PI_MEM_ADVICE_UNKNOWN,
PI_MEM_ADVICE_CUDA_SET_READ_MOSTLY = 101,
PI_MEM_ADVICE_CUDA_UNSET_READ_MOSTLY = 102,
PI_MEM_ADVICE_CUDA_SET_PREFERRED_LOCATION = 103,
PI_MEM_ADVICE_CUDA_UNSET_PREFERRED_LOCATION = 104,
PI_MEM_ADVICE_CUDA_SET_ACCESSED_BY = 105,
PI_MEM_ADVICE_CUDA_UNSET_ACCESSED_BY = 106,
PI_MEM_ADVICE_CUDA_SET_PREFERRED_LOCATION_HOST = 107,
PI_MEM_ADVICE_CUDA_UNSET_PREFERRED_LOCATION_HOST = 108,
PI_MEM_ADVICE_CUDA_SET_ACCESSED_BY_HOST = 109,
PI_MEM_ADVICE_CUDA_UNSET_ACCESSED_BY_HOST = 110,
} _pi_mem_advice;

typedef enum {
Expand Down
29 changes: 26 additions & 3 deletions sycl/plugins/cuda/pi_cuda.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -4778,9 +4778,32 @@ pi_result cuda_piextUSMEnqueueMemAdvise(pi_queue queue, const void *ptr,
event_ptr->start();
}

result = PI_CHECK_ERROR(
cuMemAdvise((CUdeviceptr)ptr, length, (CUmem_advise)advice,
queue->get_context()->get_device()->get()));
switch (advice) {
case PI_MEM_ADVICE_CUDA_SET_READ_MOSTLY:
case PI_MEM_ADVICE_CUDA_UNSET_READ_MOSTLY:
case PI_MEM_ADVICE_CUDA_SET_PREFERRED_LOCATION:
case PI_MEM_ADVICE_CUDA_UNSET_PREFERRED_LOCATION:
case PI_MEM_ADVICE_CUDA_SET_ACCESSED_BY:
case PI_MEM_ADVICE_CUDA_UNSET_ACCESSED_BY:
result = PI_CHECK_ERROR(cuMemAdvise(
(CUdeviceptr)ptr, length,
(CUmem_advise)(advice - PI_MEM_ADVICE_CUDA_SET_READ_MOSTLY + 1),
queue->get_context()->get_device()->get()));
break;
case PI_MEM_ADVICE_CUDA_SET_PREFERRED_LOCATION_HOST:
case PI_MEM_ADVICE_CUDA_UNSET_PREFERRED_LOCATION_HOST:
case PI_MEM_ADVICE_CUDA_SET_ACCESSED_BY_HOST:
case PI_MEM_ADVICE_CUDA_UNSET_ACCESSED_BY_HOST:
result = PI_CHECK_ERROR(cuMemAdvise(
(CUdeviceptr)ptr, length,
(CUmem_advise)(advice - PI_MEM_ADVICE_CUDA_SET_READ_MOSTLY + 1 -
(PI_MEM_ADVICE_CUDA_SET_PREFERRED_LOCATION_HOST -
PI_MEM_ADVICE_CUDA_SET_PREFERRED_LOCATION)),
CU_DEVICE_CPU));
break;
default:
cl::sycl::detail::pi::die("Unknown advice");
}
if (event) {
result = event_ptr->record();
*event = event_ptr.release();
Expand Down