[SYCL] Throw an exception when usm_shared_allocations aspect not supported

sycl/source/detail/usm/usm_impl.cpp

@@ -146,11 +146,8 @@ void *alignedAllocInternal(size_t Alignment, size_t Size,
   }
   if (Kind == alloc::shared &&
       !DevImpl->has(sycl::aspect::usm_shared_allocations)) {
-    // TODO:: Throw an exception to conform with the specification.
-    //  Note that many tests will have to be changed to conform with the spec
-    //  before completing this. That is, the tests will now have to expect
-    //  exceptions as a result of failed allocations in addition to nullptr
-    //  being returned depending on the reason why allocation failed.
+    throw sycl::exception(sycl::errc::feature_not_supported,
+                          "Device does not support USM shared allocations!");
   }
   void *RetVal = nullptr;
   if (Size == 0)

sycl/test-e2e/Annotated_arg_ptr/annotated_arg.cpp

@@ -75,6 +75,9 @@ MyStruct<T> operator<<(const MyStruct<T> &lhs, const MyStruct<T> &rhs) {
 
 int main() {
   queue Q;
+  if (!Q.get_device().has(aspect::usm_shared_allocations)) {
+    return 0;
+  }
 
   auto *a = malloc_shared<int>(8, Q);
   auto a_ptr = annotated_arg{a};

sycl/test-e2e/Annotated_arg_ptr/annotated_ptr.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 //

sycl/test-e2e/Annotated_usm/annotated_usm_kind.cpp

@@ -125,25 +125,27 @@ template <typename T> void testUsmKind(sycl::queue &q) {
           [&]() { return ATHost(1, q); }, [&]() { return ATHost(1, Ctx); },
           [&]() { return ATAnnotated(1, dev, Ctx, alloc::host); }});
 
-  CheckUsmKindAll(
-      alloc::shared,
-      std::tuple{
-          [&]() { return MShared(q); }, [&]() { return MShared(dev, Ctx); },
-          [&]() { return MAnnotated(dev, Ctx, alloc::shared); },
-          [&]() { return MAnnotated(dev, Ctx, properties{usm_kind_shared}); },
-          [&]() { return AShared(1, q); },
-          [&]() { return AShared(1, dev, Ctx); },
-          [&]() { return AAnnotated(1, dev, Ctx, alloc::shared); },
-          [&]() { return TShared(q); }, [&]() { return TShared(dev, Ctx); },
-          [&]() { return TAnnotated(dev, Ctx, alloc::shared); },
-          [&]() { return TAnnotated(dev, Ctx, properties{usm_kind_shared}); },
-          [&]() { return ATShared(1, q); },
-          [&]() { return ATShared(1, dev, Ctx); },
-          [&]() { return ATAnnotated(1, dev, Ctx, alloc::shared); }});
+  if (q.get_device().has(sycl::aspect::usm_shared_allocations)) {
+    CheckUsmKindAll(
+        alloc::shared,
+        std::tuple{
+            [&]() { return MShared(q); }, [&]() { return MShared(dev, Ctx); },
+            [&]() { return MAnnotated(dev, Ctx, alloc::shared); },
+            [&]() { return MAnnotated(dev, Ctx, properties{usm_kind_shared}); },
+            [&]() { return AShared(1, q); },
+            [&]() { return AShared(1, dev, Ctx); },
+            [&]() { return AAnnotated(1, dev, Ctx, alloc::shared); },
+            [&]() { return TShared(q); }, [&]() { return TShared(dev, Ctx); },
+            [&]() { return TAnnotated(dev, Ctx, alloc::shared); },
+            [&]() { return TAnnotated(dev, Ctx, properties{usm_kind_shared}); },
+            [&]() { return ATShared(1, q); },
+            [&]() { return ATShared(1, dev, Ctx); },
+            [&]() { return ATAnnotated(1, dev, Ctx, alloc::shared); }});
+  }
 }
 
 int main() {
   sycl::queue q;
   testUsmKind<int>(q);
   return 0;
-}
+}

sycl/test-e2e/Basic/group_local_memory.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 

sycl/test-e2e/Basic/large-range.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocation
 // Temporarily add explicit '-O2' to avoid GPU hang issue with O0 optimization.
 // RUN: %{build} -fno-sycl-id-queries-fit-in-int -O2 -o %t.out
 // RUN: env SYCL_PARALLEL_FOR_RANGE_ROUNDING_TRACE=1 %{run} %t.out

sycl/test-e2e/Basic/span.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 //

sycl/test-e2e/Basic/wrapped_usm_pointers.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 

sycl/test-e2e/Complex/sycl_complex_math_test.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // DEFINE: %{mathflags} = %if cl_options %{/clang:-fno-fast-math%} %else %{-fno-fast-math%}
 
 // RUN: %{build} -fsycl-device-code-split=per_kernel %{mathflags} -o %t.out

sycl/test-e2e/Complex/sycl_complex_operator_test.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // RUN: %{build} -fsycl-device-code-split=per_kernel -o %t.out
 // RUN: %{run} %t.out
 

sycl/test-e2e/Complex/sycl_complex_pow_test.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // DEFINE: %{mathflags} = %if cl_options %{/clang:-fno-fast-math%} %else %{-fno-fast-math%}
 
 // RUN: %{build} -fsycl-device-code-split=per_kernel %{mathflags} -o %t.out

sycl/test-e2e/Complex/sycl_complex_stream_test.cpp

@@ -7,6 +7,8 @@
 
 template <typename T> struct test_sycl_stream_operator {
   bool operator()(sycl::queue &Q, cmplx<T> init) {
+    if (!Q.get_device().has(sycl::aspect::usm_shared_allocations))
+      return true;
     auto *cplx_out = sycl::malloc_shared<experimental::complex<T>>(1, Q);
     cplx_out[0] = experimental::complex<T>(init.re, init.im);
 

sycl/test-e2e/DiscardEvents/discard_events_mixed_calls.cpp

@@ -37,6 +37,8 @@ static constexpr int MAX_ITER_NUM2 = 10;
 void TestHelper(sycl::queue Q,
                 const std::function<void(sycl::range<1> Range, int *Harray,
                                          sycl::buffer<int, 1> Buf)> &Function) {
+  if (!Q.get_device().has(aspect::usm_shared_allocations))
+    return;
   int *Harray = sycl::malloc_shared<int>(BUFFER_SIZE, Q);
   assert(Harray != nullptr);
   for (size_t i = 0; i < BUFFER_SIZE; ++i) {
@@ -113,36 +115,36 @@ void RunTest_USM_Accessor(sycl::queue Q) {
 }
 
 void RunTest_Accessor_USM(sycl::queue Q) {
-  TestHelper(
-      Q, [&](sycl::range<1> Range, int *Harray, sycl::buffer<int, 1> Buf) {
-        {
-          sycl::host_accessor HostAcc(Buf);
-          for (size_t i = 0; i < BUFFER_SIZE; ++i) {
-            HostAcc[i] = 0;
-          }
-        }
-
-        for (int i = 0; i < MAX_ITER_NUM1; ++i)
-          IfTrueIncrementBufferAndUSM(Q, Range, Harray, Buf, (i));
-
-        for (int i = 0; i < MAX_ITER_NUM2; ++i)
-          IfTrueIncrementUSM(Q, Range, Harray, (MAX_ITER_NUM1 + i));
-
-        Q.wait();
-
-        // check results
-        for (size_t i = 0; i < BUFFER_SIZE; ++i) {
-          int expected = MAX_ITER_NUM1 + MAX_ITER_NUM2;
-          assert(Harray[i] == expected);
-        }
-        {
-          sycl::host_accessor HostAcc(Buf, sycl::read_only);
-          for (size_t i = 0; i < BUFFER_SIZE; ++i) {
-            int expected = MAX_ITER_NUM1;
-            assert(HostAcc[i] == expected);
-          }
-        }
-      });
+  TestHelper(Q,
+             [&](sycl::range<1> Range, int *Harray, sycl::buffer<int, 1> Buf) {
+               {
+                 sycl::host_accessor HostAcc(Buf);
+                 for (size_t i = 0; i < BUFFER_SIZE; ++i) {
+                   HostAcc[i] = 0;
+                 }
+               }
+
+               for (int i = 0; i < MAX_ITER_NUM1; ++i)
+                 IfTrueIncrementBufferAndUSM(Q, Range, Harray, Buf, (i));
+
+               for (int i = 0; i < MAX_ITER_NUM2; ++i)
+                 IfTrueIncrementUSM(Q, Range, Harray, (MAX_ITER_NUM1 + i));
+
+               Q.wait();
+
+               // check results
+               for (size_t i = 0; i < BUFFER_SIZE; ++i) {
+                 int expected = MAX_ITER_NUM1 + MAX_ITER_NUM2;
+                 assert(Harray[i] == expected);
+               }
+               {
+                 sycl::host_accessor HostAcc(Buf, sycl::read_only);
+                 for (size_t i = 0; i < BUFFER_SIZE; ++i) {
+                   int expected = MAX_ITER_NUM1;
+                   assert(HostAcc[i] == expected);
+                 }
+               }
+             });
 }
 
 void RunTest_Mixed(sycl::queue Q) {

sycl/test-e2e/DiscardEvents/discard_events_usm.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // RUN: %{build} -o %t.out
 
 // RUN: env SYCL_PI_TRACE=2 %{run} %t.out &> %t.txt ; FileCheck %s --input-file %t.txt

sycl/test-e2e/DiscardEvents/discard_events_usm_ooo_queue.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // RUN: %{build} -o %t.out
 
 // RUN: env SYCL_PI_TRACE=2 %{run} %t.out &> %t.txt ; FileCheck %s --input-file %t.txt

sycl/test-e2e/DiscardEvents/invalid_event.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // Same hang as on Basic/barrier_order.cpp tracked in
 // https://github.com/intel/llvm/issues/7330.
 // UNSUPPORTED: opencl && gpu

sycl/test-e2e/GroupAlgorithm/exclusive_scan_over_group.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // CPU and ACC not yet supported:
 // Unsupported SPIR-V module SPIRV module requires unsupported capability 6400
 // REQUIRES: gpu

sycl/test-e2e/GroupAlgorithm/root_group.cpp

@@ -33,6 +33,8 @@ void testQueriesAndProperties() {
 
 void testRootGroup() {
   sycl::queue q;
+  if (!q.get_device().has(sycl::aspect::usm_shared_allocations))
+    return;
   const auto bundle =
       sycl::get_kernel_bundle<sycl::bundle_state::executable>(q.get_context());
   const auto kernel = bundle.get_kernel<class RootGroupKernel>();
@@ -66,6 +68,8 @@ void testRootGroup() {
 
 void testRootGroupFunctions() {
   sycl::queue q;
+  if (!q.get_device().has(sycl::aspect::usm_shared_allocations))
+    return;
   const auto bundle =
       sycl::get_kernel_bundle<sycl::bundle_state::executable>(q.get_context());
   const auto kernel = bundle.get_kernel<class RootGroupFunctionsKernel>();

sycl/test-e2e/InOrderEventsExt/get_last_event.cpp

@@ -34,7 +34,8 @@ int main() {
   Failed += Check(Q, "host_task", [&]() {
     return Q.submit([&](sycl::handler &CGH) { CGH.host_task([]() {}); });
   });
-
+  if (!Q.get_device().has(sycl::aspect::usm_shared_allocations))
+    return Failed;
   constexpr size_t N = 64;
   int *Data1 = sycl::malloc_shared<int>(N, Q);
   int *Data2 = sycl::malloc_shared<int>(N, Q);

sycl/test-e2e/InOrderEventsExt/set_external_event.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 

sycl/test-e2e/KernelAndProgram/disable-caching.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // This test ensures created program/kernels are not retained
 // if and only if caching is disabled.
 

sycl/test-e2e/KernelFusion/sync_two_queues_event_dep.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // For this test, complete_fusion must be supported.
 // RUN: %{build} -o %t.out
 // RUN: env SYCL_RT_WARNING_LEVEL=1 %{run} %t.out 2>&1 | FileCheck %s

sycl/test-e2e/KernelFusion/sync_usm_mem_op.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // RUN: %{build} -fsycl-embed-ir -o %t.out
 // RUN: env SYCL_RT_WARNING_LEVEL=1 %{run} %t.out 2>&1 | FileCheck %s
 

sycl/test-e2e/Reduction/reduction_internal.cpp

@@ -80,7 +80,8 @@ static void test(RedStorage &Storage, RangeTy Range) {
          cgh, Range, ext::oneapi::experimental::empty_properties_t{}, RedSycl,
          [=](auto Item, auto &Red) { Red.combine(T{1}); });
    }).wait();
-
+  if (!q.get_device().has(aspect::usm_shared_allocations))
+    return;
   auto *Result = malloc_shared<T>(1, q);
   q.submit([&](handler &cgh) {
      auto RedAcc = GetRedAcc(cgh);

sycl/test-e2e/Reduction/reduction_range_item.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 

sycl/test-e2e/Reduction/reduction_span.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 

sycl/test-e2e/Reduction/reduction_span_pack.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 //

sycl/test-e2e/Regression/exclusive-scan-char-short.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 

sycl/test-e2e/Regression/group_local_linear_id.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 

sycl/test-e2e/Regression/half_operators.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // REQUIRES: gpu
 // RUN: %{build} -fsycl-device-code-split=per_kernel -o %t.out
 // RUN: %{run} %t.out

sycl/test-e2e/Regression/pf-wg-atomic64.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // DISABLED: aspect-atomic64
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out

sycl/test-e2e/Regression/range-rounding-this-id.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // This test ensures that this_id returns the correct value
 // even when a kernel is wrapped in a range rounding kernel.
 // RUN: %{build} -o %t.out

sycl/test-e2e/Regression/reduction_64bit_atomic64.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: usm_shared_allocations
 // REQUIRES: aspect-atomic64
 // RUN: %{build} -o %t.out
 //