From 172708beb68cdaeff5a6ffba8b6f1861ad9c1eaf Mon Sep 17 00:00:00 2001 From: "aidan.belton" Date: Thu, 10 Feb 2022 11:52:28 +0000 Subject: [PATCH 01/11] add mem_advise reset and managed mem check --- sycl/include/CL/sycl/detail/pi.h | 3 ++- sycl/plugins/cuda/pi_cuda.cpp | 18 ++++++++++++++++++ 2 files changed, 20 insertions(+), 1 deletion(-) diff --git a/sycl/include/CL/sycl/detail/pi.h b/sycl/include/CL/sycl/detail/pi.h index 0e15200d19a6b..c5540b279f1bc 100644 --- a/sycl/include/CL/sycl/detail/pi.h +++ b/sycl/include/CL/sycl/detail/pi.h @@ -456,7 +456,8 @@ typedef enum { typedef enum { // Device-specific value opaque in PI API. - PI_MEM_ADVICE_UNKNOWN, + PI_MEM_ADVISE_RESET = 0, + PI_MEM_ADVICE_UNKNOWN = 1, PI_MEM_ADVICE_CUDA_SET_READ_MOSTLY = 101, PI_MEM_ADVICE_CUDA_UNSET_READ_MOSTLY = 102, PI_MEM_ADVICE_CUDA_SET_PREFERRED_LOCATION = 103, diff --git a/sycl/plugins/cuda/pi_cuda.cpp b/sycl/plugins/cuda/pi_cuda.cpp index 777e97c0f2570..6e817024cfac6 100644 --- a/sycl/plugins/cuda/pi_cuda.cpp +++ b/sycl/plugins/cuda/pi_cuda.cpp @@ -4803,6 +4803,13 @@ pi_result cuda_piextUSMEnqueueMemAdvise(pi_queue queue, const void *ptr, pi_result result = PI_SUCCESS; std::unique_ptr<_pi_event> event_ptr{nullptr}; + // Ignore mem advice if memory is not managed + unsigned int is_managed; + PI_CHECK_ERROR(cuPointerGetAttribute( + &is_managed, CU_POINTER_ATTRIBUTE_IS_MANAGED, (CUdeviceptr)ptr)); + if (!is_managed) + return PI_SUCCESS; + try { ScopedContext active(queue->get_context()); @@ -4835,6 +4842,17 @@ pi_result cuda_piextUSMEnqueueMemAdvise(pi_queue queue, const void *ptr, PI_MEM_ADVICE_CUDA_SET_PREFERRED_LOCATION)), CU_DEVICE_CPU)); break; + case PI_MEM_ADVISE_RESET: + PI_CHECK_ERROR(cuMemAdvise((CUdeviceptr)ptr, length, + CU_MEM_ADVISE_UNSET_READ_MOSTLY, + queue->get_context()->get_device()->get())); + PI_CHECK_ERROR(cuMemAdvise((CUdeviceptr)ptr, length, + CU_MEM_ADVISE_UNSET_PREFERRED_LOCATION, + queue->get_context()->get_device()->get())); + PI_CHECK_ERROR(cuMemAdvise((CUdeviceptr)ptr, length, + CU_MEM_ADVISE_UNSET_ACCESSED_BY, + queue->get_context()->get_device()->get())); + break; default: cl::sycl::detail::pi::die("Unknown advice"); } From 8ed7a36fd89779bdb5d36baea42f24aeb986bfbc Mon Sep 17 00:00:00 2001 From: "aidan.belton" Date: Mon, 9 May 2022 10:24:46 +0100 Subject: [PATCH 02/11] emit warning --- sycl/plugins/cuda/pi_cuda.cpp | 8 +++++--- 1 file changed, 5 insertions(+), 3 deletions(-) diff --git a/sycl/plugins/cuda/pi_cuda.cpp b/sycl/plugins/cuda/pi_cuda.cpp index 6e817024cfac6..9b5db62d14d48 100644 --- a/sycl/plugins/cuda/pi_cuda.cpp +++ b/sycl/plugins/cuda/pi_cuda.cpp @@ -4803,12 +4803,14 @@ pi_result cuda_piextUSMEnqueueMemAdvise(pi_queue queue, const void *ptr, pi_result result = PI_SUCCESS; std::unique_ptr<_pi_event> event_ptr{nullptr}; - // Ignore mem advice if memory is not managed unsigned int is_managed; PI_CHECK_ERROR(cuPointerGetAttribute( &is_managed, CU_POINTER_ATTRIBUTE_IS_MANAGED, (CUdeviceptr)ptr)); - if (!is_managed) - return PI_SUCCESS; + if (!is_managed) { + setErrorMessage("Prefetch hint ignored as mem advise only works with USM", + PI_SUCCESS); + return PI_PLUGIN_SPECIFIC_ERROR; + } try { ScopedContext active(queue->get_context()); From 475b1686837a577a69b1a8f411cde03be7ebb482 Mon Sep 17 00:00:00 2001 From: "aidan.belton" Date: Tue, 28 Jun 2022 10:28:21 +0100 Subject: [PATCH 03/11] remove hint --- sycl/plugins/cuda/pi_cuda.cpp | 9 --------- 1 file changed, 9 deletions(-) diff --git a/sycl/plugins/cuda/pi_cuda.cpp b/sycl/plugins/cuda/pi_cuda.cpp index 46851a263c180..0ecf2075357ac 100644 --- a/sycl/plugins/cuda/pi_cuda.cpp +++ b/sycl/plugins/cuda/pi_cuda.cpp @@ -5112,15 +5112,6 @@ pi_result cuda_piextUSMEnqueueMemAdvise(pi_queue queue, const void *ptr, pi_result result = PI_SUCCESS; std::unique_ptr<_pi_event> event_ptr{nullptr}; - unsigned int is_managed; - PI_CHECK_ERROR(cuPointerGetAttribute( - &is_managed, CU_POINTER_ATTRIBUTE_IS_MANAGED, (CUdeviceptr)ptr)); - if (!is_managed) { - setErrorMessage("Prefetch hint ignored as mem advise only works with USM", - PI_SUCCESS); - return PI_PLUGIN_SPECIFIC_ERROR; - } - try { ScopedContext active(queue->get_context()); From 958f197af07e0acb3fa88c0dce2ce190dda00eae Mon Sep 17 00:00:00 2001 From: AidanBeltonS <87009434+AidanBeltonS@users.noreply.github.com> Date: Mon, 18 Jul 2022 15:02:39 +0100 Subject: [PATCH 04/11] Rename PI_MEM_ADVISE_RESET to PI_MEM_ADVICE_RESET Co-authored-by: Steffen Larsen --- sycl/include/CL/sycl/detail/pi.h | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/sycl/include/CL/sycl/detail/pi.h b/sycl/include/CL/sycl/detail/pi.h index 44d4b92a4cb66..b95ff4d744f35 100644 --- a/sycl/include/CL/sycl/detail/pi.h +++ b/sycl/include/CL/sycl/detail/pi.h @@ -404,7 +404,7 @@ typedef enum { typedef enum { // Device-specific value opaque in PI API. - PI_MEM_ADVISE_RESET = 0, + PI_MEM_ADVICE_RESET = 0, PI_MEM_ADVICE_UNKNOWN = 1, PI_MEM_ADVICE_CUDA_SET_READ_MOSTLY = 101, PI_MEM_ADVICE_CUDA_UNSET_READ_MOSTLY = 102, From 753ed95d5f5e7a2a3c811aeeb45e0feed116a74f Mon Sep 17 00:00:00 2001 From: "aidan.belton" Date: Mon, 18 Jul 2022 15:15:11 +0100 Subject: [PATCH 05/11] Check for concurrent managed access with PI_MEM_ADVISE_RESET --- sycl/plugins/cuda/pi_cuda.cpp | 3 ++- 1 file changed, 2 insertions(+), 1 deletion(-) diff --git a/sycl/plugins/cuda/pi_cuda.cpp b/sycl/plugins/cuda/pi_cuda.cpp index 0ecf2075357ac..008cb3497139b 100644 --- a/sycl/plugins/cuda/pi_cuda.cpp +++ b/sycl/plugins/cuda/pi_cuda.cpp @@ -5095,7 +5095,8 @@ pi_result cuda_piextUSMEnqueueMemAdvise(pi_queue queue, const void *ptr, if (advice == PI_MEM_ADVICE_CUDA_SET_PREFERRED_LOCATION || advice == PI_MEM_ADVICE_CUDA_UNSET_PREFERRED_LOCATION || advice == PI_MEM_ADVICE_CUDA_SET_ACCESSED_BY || - advice == PI_MEM_ADVICE_CUDA_UNSET_ACCESSED_BY) { + advice == PI_MEM_ADVICE_CUDA_UNSET_ACCESSED_BY || + advice == PI_MEM_ADVISE_RESET) { pi_device device = queue->get_context()->get_device(); if (!getAttribute(device, CU_DEVICE_ATTRIBUTE_CONCURRENT_MANAGED_ACCESS)) { setErrorMessage("Mem advise ignored as device does not support " From 2c1062b61fd3d5f8d994b68e9aaaa5be1e82dc25 Mon Sep 17 00:00:00 2001 From: "aidan.belton" Date: Mon, 18 Jul 2022 15:29:22 +0100 Subject: [PATCH 06/11] Increment SYCL_DEV_ABI_VERSION --- sycl/CMakeLists.txt | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/sycl/CMakeLists.txt b/sycl/CMakeLists.txt index 64820ebfbae47..779caff275ae6 100644 --- a/sycl/CMakeLists.txt +++ b/sycl/CMakeLists.txt @@ -28,7 +28,7 @@ include(SYCLUtils) set(SYCL_MAJOR_VERSION 5) set(SYCL_MINOR_VERSION 7) set(SYCL_PATCH_VERSION 0) -set(SYCL_DEV_ABI_VERSION 0) +set(SYCL_DEV_ABI_VERSION 1) if (SYCL_ADD_DEV_VERSION_POSTFIX) set(SYCL_VERSION_POSTFIX "-${SYCL_DEV_ABI_VERSION}") endif() From 45655bc1b2dae2efbaf9b10cee0bf3843e32483d Mon Sep 17 00:00:00 2001 From: "aidan.belton" Date: Mon, 18 Jul 2022 15:33:52 +0100 Subject: [PATCH 07/11] Change PI_MEM_ADVICE_RESET value --- sycl/include/CL/sycl/detail/pi.h | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/sycl/include/CL/sycl/detail/pi.h b/sycl/include/CL/sycl/detail/pi.h index 44d4b92a4cb66..13673c16cc6b6 100644 --- a/sycl/include/CL/sycl/detail/pi.h +++ b/sycl/include/CL/sycl/detail/pi.h @@ -405,7 +405,6 @@ typedef enum { typedef enum { // Device-specific value opaque in PI API. PI_MEM_ADVISE_RESET = 0, - PI_MEM_ADVICE_UNKNOWN = 1, PI_MEM_ADVICE_CUDA_SET_READ_MOSTLY = 101, PI_MEM_ADVICE_CUDA_UNSET_READ_MOSTLY = 102, PI_MEM_ADVICE_CUDA_SET_PREFERRED_LOCATION = 103, @@ -416,6 +415,7 @@ typedef enum { 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_UNKNOWN = 999, } _pi_mem_advice; typedef enum { From d61e28c6e3eef349faf5a8cdfe503ff9ed8c9eb8 Mon Sep 17 00:00:00 2001 From: "aidan.belton" Date: Mon, 18 Jul 2022 15:45:50 +0100 Subject: [PATCH 08/11] Transfer changes to new pi.h directory --- sycl/include/sycl/detail/pi.h | 3 ++- 1 file changed, 2 insertions(+), 1 deletion(-) diff --git a/sycl/include/sycl/detail/pi.h b/sycl/include/sycl/detail/pi.h index 30e65a9310a3d..0d5903707797d 100644 --- a/sycl/include/sycl/detail/pi.h +++ b/sycl/include/sycl/detail/pi.h @@ -404,7 +404,7 @@ typedef enum { typedef enum { // Device-specific value opaque in PI API. - PI_MEM_ADVICE_UNKNOWN, + PI_MEM_ADVICE_RESET = 0, PI_MEM_ADVICE_CUDA_SET_READ_MOSTLY = 101, PI_MEM_ADVICE_CUDA_UNSET_READ_MOSTLY = 102, PI_MEM_ADVICE_CUDA_SET_PREFERRED_LOCATION = 103, @@ -415,6 +415,7 @@ typedef enum { 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_UNKNOWN = 999, } _pi_mem_advice; typedef enum { From c9143c25983b9e5fbc1208d790aace3f0f39f105 Mon Sep 17 00:00:00 2001 From: "aidan.belton" Date: Mon, 18 Jul 2022 16:00:19 +0100 Subject: [PATCH 09/11] Fix typo --- sycl/plugins/cuda/pi_cuda.cpp | 4 ++-- 1 file changed, 2 insertions(+), 2 deletions(-) diff --git a/sycl/plugins/cuda/pi_cuda.cpp b/sycl/plugins/cuda/pi_cuda.cpp index 71fd2fee4433a..bce09721e5b1f 100644 --- a/sycl/plugins/cuda/pi_cuda.cpp +++ b/sycl/plugins/cuda/pi_cuda.cpp @@ -5100,7 +5100,7 @@ pi_result cuda_piextUSMEnqueueMemAdvise(pi_queue queue, const void *ptr, advice == PI_MEM_ADVICE_CUDA_UNSET_PREFERRED_LOCATION || advice == PI_MEM_ADVICE_CUDA_SET_ACCESSED_BY || advice == PI_MEM_ADVICE_CUDA_UNSET_ACCESSED_BY || - advice == PI_MEM_ADVISE_RESET) { + advice == PI_MEM_ADVICE_RESET) { pi_device device = queue->get_context()->get_device(); if (!getAttribute(device, CU_DEVICE_ATTRIBUTE_CONCURRENT_MANAGED_ACCESS)) { setErrorMessage("Mem advise ignored as device does not support " @@ -5149,7 +5149,7 @@ pi_result cuda_piextUSMEnqueueMemAdvise(pi_queue queue, const void *ptr, PI_MEM_ADVICE_CUDA_SET_PREFERRED_LOCATION)), CU_DEVICE_CPU)); break; - case PI_MEM_ADVISE_RESET: + case PI_MEM_ADVICE_RESET: PI_CHECK_ERROR(cuMemAdvise((CUdeviceptr)ptr, length, CU_MEM_ADVISE_UNSET_READ_MOSTLY, queue->get_context()->get_device()->get())); From bfbfbb02bf028d005afd6c19751ef0239e89ee2b Mon Sep 17 00:00:00 2001 From: "aidan.belton" Date: Fri, 29 Jul 2022 09:29:57 +0100 Subject: [PATCH 10/11] Incremement PI version --- sycl/include/sycl/detail/pi.h | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/sycl/include/sycl/detail/pi.h b/sycl/include/sycl/detail/pi.h index 0d5903707797d..47bdf5462bffc 100644 --- a/sycl/include/sycl/detail/pi.h +++ b/sycl/include/sycl/detail/pi.h @@ -49,7 +49,7 @@ // NOTE that this results in a changed API for `piProgramGetBuildInfo`. #define _PI_H_VERSION_MAJOR 9 -#define _PI_H_VERSION_MINOR 11 +#define _PI_H_VERSION_MINOR 12 #define _PI_STRING_HELPER(a) #a #define _PI_CONCAT(a, b) _PI_STRING_HELPER(a.b) From 21cd50ba52b179aa53659226a7e97a1aa8d53dc0 Mon Sep 17 00:00:00 2001 From: "aidan.belton" Date: Fri, 29 Jul 2022 09:40:04 +0100 Subject: [PATCH 11/11] Increment PI major version and add comment --- sycl/include/sycl/detail/pi.h | 4 +++- 1 file changed, 3 insertions(+), 1 deletion(-) diff --git a/sycl/include/sycl/detail/pi.h b/sycl/include/sycl/detail/pi.h index 47bdf5462bffc..67094702273d6 100644 --- a/sycl/include/sycl/detail/pi.h +++ b/sycl/include/sycl/detail/pi.h @@ -47,8 +47,10 @@ // 8.10 Added new optional device argument to piextQueueCreateWithNativeHandle // 9.11 Use values of OpenCL enums directly, rather than including ``; // NOTE that this results in a changed API for `piProgramGetBuildInfo`. +// 10.12 Change enum value PI_MEM_ADVICE_UNKNOWN from 0 to 999, and set enum +// PI_MEM_ADVISE_RESET to 0. -#define _PI_H_VERSION_MAJOR 9 +#define _PI_H_VERSION_MAJOR 10 #define _PI_H_VERSION_MINOR 12 #define _PI_STRING_HELPER(a) #a