diff --git a/sycl/test-e2e/DiscardEvents/invalid_event.cpp b/sycl/test-e2e/DiscardEvents/invalid_event.cpp index 4fc0ee0f1495b..17c6a492a10c2 100644 --- a/sycl/test-e2e/DiscardEvents/invalid_event.cpp +++ b/sycl/test-e2e/DiscardEvents/invalid_event.cpp @@ -2,7 +2,6 @@ // https://github.com/intel/llvm/issues/7330. // UNSUPPORTED: opencl && gpu // RUN: %{build} -o %t.out -// REQUIRES: aspect-usm_shared_allocations // RUN: %{run} %t.out // The test checks that each PI call to the queue returns a discarded event @@ -19,9 +18,9 @@ void QueueAPIsReturnDiscardedEvent(sycl::queue Q) { sycl::range<1> range(BUFFER_SIZE); auto Dev = Q.get_device(); - int *x = sycl::malloc_shared(BUFFER_SIZE, Q); + int *x = sycl::malloc_device(BUFFER_SIZE, Q); assert(x != nullptr); - int *y = sycl::malloc_shared(BUFFER_SIZE, Q); + int *y = sycl::malloc_device(BUFFER_SIZE, Q); assert(y != nullptr); sycl::event DiscardedEvent; diff --git a/sycl/test-e2e/GroupAlgorithm/root_group.cpp b/sycl/test-e2e/GroupAlgorithm/root_group.cpp index 584ecfd22fe0b..e346ea142b759 100644 --- a/sycl/test-e2e/GroupAlgorithm/root_group.cpp +++ b/sycl/test-e2e/GroupAlgorithm/root_group.cpp @@ -41,10 +41,10 @@ void testRootGroup() { max_num_work_group_sync>(q); const auto props = sycl::ext::oneapi::experimental::properties{ sycl::ext::oneapi::experimental::use_root_sync}; - sycl::buffer data_buf{sycl::range{maxWGs * WorkGroupSize}}; + sycl::buffer dataBuf{sycl::range{maxWGs * WorkGroupSize}}; const auto range = sycl::nd_range<1>{maxWGs * WorkGroupSize, WorkGroupSize}; q.submit([&](sycl::handler &h) { - sycl::accessor data{data_buf, h}; + sycl::accessor data{dataBuf, h}; h.parallel_for( range, props, [=](sycl::nd_item<1> it) { auto root = it.ext_oneapi_get_root_group(); @@ -59,7 +59,7 @@ void testRootGroup() { data[root.get_local_id()] = sum; }); }); - sycl::host_accessor data{data_buf}; + sycl::host_accessor data{dataBuf}; const int workItemCount = static_cast(range.get_global_range().size()); for (int i = 0; i < workItemCount; i++) { assert(data[i] == (workItemCount - 1)); @@ -78,10 +78,10 @@ void testRootGroupFunctions() { sycl::ext::oneapi::experimental::use_root_sync}; constexpr int testCount = 10; - sycl::buffer testResults_buf{sycl::range{testCount}}; + sycl::buffer testResultsBuf{sycl::range{testCount}}; const auto range = sycl::nd_range<1>{maxWGs * WorkGroupSize, WorkGroupSize}; q.submit([&](sycl::handler &h) { - sycl::accessor testResults{testResults_buf, h}; + sycl::accessor testResults{testResultsBuf, h}; h.parallel_for( range, props, [=](sycl::nd_item<1> it) { const auto root = it.ext_oneapi_get_root_group(); @@ -113,7 +113,7 @@ void testRootGroupFunctions() { } }); }); - sycl::host_accessor testResults{testResults_buf}; + sycl::host_accessor testResults{testResultsBuf}; for (int i = 0; i < testCount; i++) { assert(testResults[i]); } diff --git a/sycl/test-e2e/InOrderEventsExt/set_external_event.cpp b/sycl/test-e2e/InOrderEventsExt/set_external_event.cpp index 4f2a409fa3d46..61987d5b5b3cb 100644 --- a/sycl/test-e2e/InOrderEventsExt/set_external_event.cpp +++ b/sycl/test-e2e/InOrderEventsExt/set_external_event.cpp @@ -15,8 +15,8 @@ int main() { sycl::queue Q1{Ctx, Dev, {sycl::property::queue::in_order{}}}; sycl::queue Q2{Ctx, Dev, {sycl::property::queue::in_order{}}}; - sycl::buffer DevDatabuf{sycl::range{N}}; - sycl::accessor DevData{DevDatabuf}; + sycl::buffer DevDataBuf{sycl::range{N}}; + sycl::accessor DevData{DevDataBuf}; int *HostData = (int *)malloc(N * sizeof(int) * 10); for (size_t I = 0; I < 10; ++I) { @@ -52,6 +52,6 @@ int main() { } } } - + free(HostData); return Failures; } diff --git a/sycl/test-e2e/Reduction/reduction_internal.cpp b/sycl/test-e2e/Reduction/reduction_internal.cpp index e16f8d750a154..39c64932802da 100644 --- a/sycl/test-e2e/Reduction/reduction_internal.cpp +++ b/sycl/test-e2e/Reduction/reduction_internal.cpp @@ -80,13 +80,13 @@ 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(); - sycl::buffer Result_buf{sycl::range{1}}; + sycl::buffer ResultBuf{sycl::range{1}}; q.submit([&](handler &cgh) { - sycl::accessor Result{Result_buf, cgh}; + sycl::accessor Result{ResultBuf, cgh}; auto RedAcc = GetRedAcc(cgh); cgh.single_task([=]() { Result[0] = RedAcc[0]; }); }); - sycl::host_accessor Result{Result_buf}; + sycl::host_accessor Result{ResultBuf}; auto N = get_global_range(Range).size(); int Expected = InitToIdentity ? N : Init + N; #if defined(__PRETTY_FUNCTION__) diff --git a/sycl/test-e2e/Reduction/reduction_range_item.cpp b/sycl/test-e2e/Reduction/reduction_range_item.cpp index 485959860456d..06d4f33280f06 100644 --- a/sycl/test-e2e/Reduction/reduction_range_item.cpp +++ b/sycl/test-e2e/Reduction/reduction_range_item.cpp @@ -3,15 +3,14 @@ // Windows doesn't yet have full shutdown(). // UNSUPPORTED: ze_debug && windows -// REQUIRES: aspect-usm_shared_allocations #include using namespace sycl; int main() { queue q; - auto *RedMem = malloc_shared(1, q); - auto *Success = malloc_shared(1, q); + auto *RedMem = malloc_device(1, q); + auto *Success = malloc_device(1, q); *Success = true; *RedMem = 0; @@ -24,11 +23,15 @@ int main() { *Success = false; }) .wait(); - - assert(*RedMem == 7); - assert(*Success); - - *RedMem = 0; + int RedMemHost; + bool SuccessHost; + q.memcpy(&RedMemHost, RedMem, sizeof(int)).wait(); + q.memcpy(&SuccessHost, Success, sizeof(bool)).wait(); + assert(RedMemHost == 7); + assert(SuccessHost); + + RedMemHost = 0; + q.memcpy(RedMem, &RedMemHost, sizeof(int)).wait(); q.parallel_for(range<2>{1030, 7}, reduction(RedMem, std::plus{}), [=](item<2> Item, auto &Red) { Red += 1; @@ -44,8 +47,10 @@ int main() { }) .wait(); - assert(*RedMem == 1030 * 7); - assert(*Success); + q.memcpy(&RedMemHost, RedMem, sizeof(int)).wait(); + q.memcpy(&SuccessHost, Success, sizeof(bool)).wait(); + assert(RedMemHost == 1030 * 7); + assert(SuccessHost); free(RedMem, q); free(Success, q); diff --git a/sycl/test-e2e/Reduction/reduction_span.cpp b/sycl/test-e2e/Reduction/reduction_span.cpp index 63f5ec1ece9be..23ee61e805f9b 100644 --- a/sycl/test-e2e/Reduction/reduction_span.cpp +++ b/sycl/test-e2e/Reduction/reduction_span.cpp @@ -3,7 +3,6 @@ // Windows doesn't yet have full shutdown(). // UNSUPPORTED: ze_debug && windows -// REQUIRES: aspect-usm_shared_allocations // This test performs basic checks of reductions initialized with a sycl::span #include @@ -44,7 +43,7 @@ template (N, Q); + T *Output = malloc_device(N, Q); Q.parallel_for(range<1>{N}, [=](id<1> I) { Output[I] = Identity; }).wait(); // Perform generalized "histogram" with N bins @@ -70,11 +69,13 @@ void test(queue Q, Range Rng, T Identity, T Value) { } bool Passed = true; + T OutputHost[N]; + Q.memcpy(OutputHost, Output, N * sizeof(T)).wait(); for (size_t I = 0; I < N; ++I) { if (I < Size % N) { - Passed &= (Output[I] == Expected); + Passed &= (OutputHost[I] == Expected); } else { - Passed &= (Output[I] == ExpectedRemainder); + Passed &= (OutputHost[I] == ExpectedRemainder); } } diff --git a/sycl/test-e2e/Reduction/reduction_span_pack.cpp b/sycl/test-e2e/Reduction/reduction_span_pack.cpp index 706dca1363dd3..b6e118c081cb0 100644 --- a/sycl/test-e2e/Reduction/reduction_span_pack.cpp +++ b/sycl/test-e2e/Reduction/reduction_span_pack.cpp @@ -6,7 +6,6 @@ // Windows doesn't yet have full shutdown(). // UNSUPPORTED: ze_debug && windows -// REQUIRES: aspect-usm_shared_allocations // This test performs basic checks of reductions initialized with a pack // containing at least one sycl::span @@ -49,9 +48,9 @@ template (1, Q); + int *Sum = malloc_device(1, Q); Q.single_task([=]() { *Sum = 0; }).wait(); - T *Output = malloc_shared(N, Q); + T *Output = malloc_device(N, Q); Q.parallel_for(range<1>{N}, [=](id<1> I) { Output[I] = Identity; }).wait(); // Perform generalized "histogram" with N bins @@ -81,11 +80,13 @@ void test1(queue Q, Range Rng, T Identity, T Value) { } bool Passed = true; + T OutputHost[N]; + Q.memcpy(OutputHost, Output, N * sizeof(T)).wait(); for (size_t I = 0; I < N; ++I) { if (I < Size % N) { - Passed &= (Output[I] == Expected); + Passed &= (OutputHost[I] == Expected); } else { - Passed &= (Output[I] == ExpectedRemainder); + Passed &= (OutputHost[I] == ExpectedRemainder); } } Passed &= (*Sum == Size); @@ -101,9 +102,9 @@ template (N, Q); + int *Output1 = malloc_device(N, Q); Q.parallel_for(range<1>{N}, [=](id<1> I) { Output1[I] = 0; }).wait(); - T *Output2 = malloc_shared(N, Q); + T *Output2 = malloc_device(N, Q); Q.parallel_for(range<1>{N}, [=](id<1> I) { Output2[I] = Identity; }).wait(); // Perform generalized "histogram" with N bins @@ -121,7 +122,10 @@ void test2(queue Q, Range Rng, T Identity, T Value) { } else /*if (SubmissionMode == submission_mode::queue) */ { Q.parallel_for(Rng, Redu1, Redu2, Kern).wait(); } - + int Output1Host[N]; + T Output2Host[N]; + Q.memcpy(Output1Host, Output1, N * sizeof(int)).wait(); + Q.memcpy(Output2Host, Output2, N * sizeof(T)).wait(); size_t Size = getLinearSize(Rng); bool Passed = true; // Span1 @@ -132,12 +136,11 @@ void test2(queue Q, Range Rng, T Identity, T Value) { ExpectedRemainder = Expected; Expected += 1; } - for (size_t I = 0; I < N; ++I) { if (I < Size % N) { - Passed &= (Output1[I] == Expected); + Passed &= (Output1Host[I] == Expected); } else { - Passed &= (Output1[I] == ExpectedRemainder); + Passed &= (Output1Host[I] == ExpectedRemainder); } } } @@ -153,9 +156,9 @@ void test2(queue Q, Range Rng, T Identity, T Value) { for (size_t I = 0; I < N; ++I) { if (I < Size % N) { - Passed &= (Output2[I] == Expected); + Passed &= (Output2Host[I] == Expected); } else { - Passed &= (Output2[I] == ExpectedRemainder); + Passed &= (Output2Host[I] == ExpectedRemainder); } } } diff --git a/sycl/test-e2e/Regression/reduction_64bit_atomic64.cpp b/sycl/test-e2e/Regression/reduction_64bit_atomic64.cpp index 11affac9df7b9..c2d0d3e84ca7c 100644 --- a/sycl/test-e2e/Regression/reduction_64bit_atomic64.cpp +++ b/sycl/test-e2e/Regression/reduction_64bit_atomic64.cpp @@ -2,7 +2,6 @@ // RUN: %{build} -o %t.out // // RUN: %{run} %t.out -// REQUIRES: aspect-usm_shared_allocations // Tests that a previously known case for reduction doesn't cause a requirement // for atomic64. // TODO: When aspect requirements are added to testing, this test could be set @@ -19,7 +18,7 @@ using namespace sycl; int main() { queue Q; - long long *Out = malloc_shared(1, Q); + long long *Out = malloc_device(1, Q); // Case 1: nd_range reduction with 64-bit integer and either sycl::plus, // sycl::minimum or sycl::maximum. group_reduce_and_atomic_cross_wg strategy