[SYCL][CUDA] Make plugin specific error return an error (intel#10626)

The `UR_RESULT_ADAPTER_SPECIFIC_ERROR` was not returning an error to the
SYCL RT which meant all errors were treated as warnings and ignored
unless `SYCL_RT_WARNING_LEVEL` is set to geq 2. This changes things so
the adapter specific error is now reported as such, meaning all uses
`UR_RESULT_ADAPTER_SPECIFIC_ERROR` meant as warnings are now caught as
errors.

---------

Co-authored-by: Hugh Delaney <[email protected]>

Author: 2 people

sycl/plugins/unified_runtime/pi2ur.hpp

@@ -932,12 +932,14 @@ inline pi_result piPluginGetLastError(char **Message) {
   // reference for the urAdapterGetLastError call, then release it.
   ur_adapter_handle_t Adapter;
   urAdapterGet(1, &Adapter, nullptr);
+  // FIXME: ErrorCode should store a native error, but these are not being used
+  // in CUDA adapter at the moment
   int32_t ErrorCode;
-  urAdapterGetLastError(Adapter, const_cast<const char **>(Message),
-                        &ErrorCode);
+  ur_result_t Res = urAdapterGetLastError(
+      Adapter, const_cast<const char **>(Message), &ErrorCode);
   urAdapterRelease(Adapter);
 
-  return PI_SUCCESS;
+  return ur2piResult(Res);
 }
 
 inline pi_result piDeviceGetInfo(pi_device Device, pi_device_info ParamName,

sycl/plugins/unified_runtime/ur/adapters/cuda/adapter.cpp

@@ -64,9 +64,9 @@ UR_APIEXPORT ur_result_t UR_APICALL urAdapterRelease(ur_adapter_handle_t) {
 
 UR_APIEXPORT ur_result_t UR_APICALL urAdapterGetLastError(
     ur_adapter_handle_t, const char **ppMessage, int32_t *pError) {
+  std::ignore = pError;
   *ppMessage = ErrorMessage;
-  *pError = ErrorMessageCode;
-  return UR_RESULT_SUCCESS;
+  return ErrorMessageCode;
 }
 
 UR_APIEXPORT ur_result_t UR_APICALL urAdapterGetInfo(ur_adapter_handle_t,

sycl/test-e2e/Plugin/cuda-max-local-mem-size.cpp

@@ -0,0 +1,30 @@
+// REQUIRES: cuda
+
+// RUN: %{build} -o %t.out
+// RUN: not %{run} SYCL_PI_CUDA_MAX_LOCAL_MEM_SIZE=0 %t.out 2>&1 | FileCheck --check-prefixes=CHECK-ZERO %s
+// RUN: not %{run} SYCL_PI_CUDA_MAX_LOCAL_MEM_SIZE=100000000 %t.out 2>&1 | FileCheck --check-prefixes=CHECK-OVERALLOCATE %s
+
+//==---------------------- cuda-max-local-mem-size.cpp --------------------===//
+//==--- SYCL test to test SYCL_PI_CUDA_MAX_LOCAL_MEM_SIZE env var----------===//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+//
+//===----------------------------------------------------------------------===//
+
+#include <sycl/sycl.hpp>
+
+int main() {
+  sycl::queue Q{};
+  auto LocalSize =
+      Q.get_device().get_info<sycl::info::device::local_mem_size>();
+  Q.submit([&](sycl::handler &cgh) {
+     auto LocalAcc = sycl::local_accessor<float>(LocalSize + 1, cgh);
+     cgh.parallel_for(sycl::nd_range<1>{32, 32}, [=](sycl::nd_item<1> idx) {
+       LocalAcc[idx.get_global_linear_id()] *= 2;
+     });
+   }).wait();
+  // CHECK-ZERO: Local memory for kernel exceeds the amount requested using SYCL_PI_CUDA_MAX_LOCAL_MEM_SIZE
+  // CHECK-OVERALLOCATE: Too much local memory allocated for device
+}