Skip to content

[SYCL] [E2E] Fix free funcs tests #17742

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 2 commits into from
Apr 9, 2025
Merged

[SYCL] [E2E] Fix free funcs tests #17742

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
2 commits
Select commit Hold shift + click to select a range
460d07f
[SYCL] [E2E] Fix free funcs tests
jchlanda Mar 31, 2025
496a92b
Merge remote-tracking branch 'upstream/sycl' into jakub/free_funcs
jchlanda Apr 9, 2025
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
95 changes: 26 additions & 69 deletions sycl/test-e2e/Basic/free_function_queries/free_function_queries.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -1,10 +1,7 @@
// FIXME: Investigate OS-agnostic failures
// UNSUPPORTED: true
// RUN: %{build} -o %t.out
// RUN: %{run} %t.out

// UNSUPPORTED: windows
// The failure is caused by intel/llvm#5213
// With awkward sizes (such as 1030) make sure that there is no range rounding,
// so `get_global_id` returns correct values.
// RUN: env SYCL_DISABLE_PARALLEL_FOR_RANGE_ROUNDING=1 %{run} %t.out

//==- free_function_queries.cpp - SYCL free function queries test -=//
//
Expand All @@ -29,7 +26,7 @@ int main() {

{
constexpr int checks_number{3};
int results[checks_number]{};
int results[checks_number]{41, 42, 43};
{
sycl::buffer<int> buf(data, sycl::range<1>(n));
sycl::buffer<int> results_buf(results, sycl::range<1>(checks_number));
Expand All @@ -42,12 +39,12 @@ int main() {
sycl::access::target::device>
results_acc(results_buf.get_access<sycl::access::mode::write>(cgh));
cgh.parallel_for<class IdTest>(n, [=](sycl::id<1> i) {
auto that_id = sycl::ext::oneapi::experimental::this_id<1>();
results_acc[0] = that_id == i;
auto that_id = sycl::ext::oneapi::this_work_item::get_nd_item<1>();
results_acc[0] = that_id.get_global_id() == i;

auto that_item = sycl::ext::oneapi::experimental::this_item<1>();
results_acc[1] = that_item.get_id() == i;
results_acc[2] = that_item.get_range() == sycl::range<1>(n);
auto that_item = sycl::ext::oneapi::this_work_item::get_nd_item<1>();
results_acc[1] = that_item.get_global_id() == i;
results_acc[2] = that_item.get_global_range() == sycl::range<1>(n);
acc[i]++;
});
});
Expand All @@ -63,7 +60,7 @@ int main() {

{
constexpr int checks_number{2};
int results[checks_number]{};
int results[checks_number]{41, 42};
{
sycl::buffer<int> buf(data, sycl::range<1>(n));
sycl::buffer<int> results_buf(results, sycl::range<1>(checks_number));
Expand All @@ -78,10 +75,10 @@ int main() {
cgh.parallel_for<class ItemTest>(n, [=](auto i) {
static_assert(std::is_same<decltype(i), sycl::item<1>>::value,
"lambda arg type is unexpected");
auto that_id = sycl::ext::oneapi::experimental::this_id<1>();
results_acc[0] = i.get_id() == that_id;
auto that_item = sycl::ext::oneapi::experimental::this_item<1>();
results_acc[1] = i == that_item;
auto that_id = sycl::ext::oneapi::this_work_item::get_nd_item<1>();
results_acc[0] = that_id.get_global_id() == i;
auto that_item = sycl::ext::oneapi::this_work_item::get_nd_item<1>();
results_acc[1] = that_item.get_global_id() == i;
acc[i]++;
});
});
Expand All @@ -96,8 +93,11 @@ int main() {
}

{
// Make sure that we ignore global offset warning.
#pragma GCC diagnostic push
#pragma GCC diagnostic ignored "-Wdeprecated-declarations"
constexpr int checks_number{2};
int results[checks_number]{};
int results[checks_number]{41, 42};
{
sycl::buffer<int> buf(data, sycl::range<1>(n));
sycl::buffer<int> results_buf(results, sycl::range<1>(checks_number));
Expand All @@ -112,11 +112,13 @@ int main() {
results_acc(results_buf.get_access<sycl::access::mode::write>(cgh));
cgh.parallel_for<class ItemOffsetTest>(
sycl::range<1>{n}, offset, [=](sycl::item<1, true> i) {
auto that_id = sycl::ext::oneapi::experimental::this_id<1>();
results_acc[0] = i.get_id() == that_id;
auto that_item = sycl::ext::oneapi::experimental::this_item<1>();
results_acc[1] = i == that_item;
acc[that_item.get_linear_id()]++;
auto that_id =
sycl::ext::oneapi::this_work_item::get_nd_item<1>();
results_acc[0] = i.get_id() == that_id.get_global_id();
auto that_item =
sycl::ext::oneapi::this_work_item::get_nd_item<1>();
results_acc[1] = i == that_item.get_global_id();
acc[that_item.get_global_linear_id()]++;
});
});
}
Expand All @@ -127,51 +129,6 @@ int main() {
for (auto val : results) {
assert(val == 1);
}
}

{
constexpr int checks_number{5};
int results[checks_number]{};
{
sycl::buffer<int> buf(data, sycl::range<1>(n));
sycl::buffer<int> results_buf(results, sycl::range<1>(checks_number));
sycl::queue q;
sycl::nd_range<1> NDR(sycl::range<1>{n}, sycl::range<1>{2});
q.submit([&](sycl::handler &cgh) {
sycl::accessor<int, 1, sycl::access::mode::write,
sycl::access::target::device>
acc(buf.get_access<sycl::access::mode::write>(cgh));
sycl::accessor<int, 1, sycl::access::mode::write,
sycl::access::target::device>
results_acc(results_buf.get_access<sycl::access::mode::write>(cgh));
cgh.parallel_for<class NdItemTest>(NDR, [=](auto nd_i) {
static_assert(std::is_same<decltype(nd_i), sycl::nd_item<1>>::value,
"lambda arg type is unexpected");
auto that_nd_item =
sycl::ext::oneapi::this_work_item::get_nd_item<1>();
results_acc[0] = that_nd_item == nd_i;
auto nd_item_group = that_nd_item.get_group();
results_acc[1] =
nd_item_group ==
sycl::ext::oneapi::this_work_item::get_work_group<1>();
auto nd_item_id = that_nd_item.get_global_id();
results_acc[2] =
nd_item_id == sycl::ext::oneapi::experimental::this_id<1>();
auto that_item = sycl::ext::oneapi::experimental::this_item<1>();
results_acc[3] = nd_item_id == that_item.get_id();
results_acc[4] =
that_nd_item.get_global_range() == that_item.get_range();

acc[that_nd_item.get_global_id(0)]++;
});
});
}
++counter;
for (int i = 0; i < n; i++) {
assert(data[i] == counter);
}
for (auto val : results) {
assert(val == 1);
}
#pragma GCC diagnostic pop
}
}
14 changes: 1 addition & 13 deletions sycl/test-e2e/Basic/free_function_queries/free_function_queries_sub_group.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -1,14 +1,6 @@
// FIXME: Investigate OS-agnostic failures
// UNSUPPORTED: true
// UNSUPPORTED: cuda || hip
// CUDA and HIP compilation and runtime do not yet support sub-groups.
//
// RUN: %{build} -o %t.out
// RUN: %{run} %t.out

// UNSUPPORTED: windows
// The failure is caused by intel/llvm#5213

//==- free_function_queries_sub_group.cpp - SYCL free queries for sub group -=//
//
// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
Expand All @@ -31,15 +23,11 @@ int main() {
int counter{0};
{
constexpr int checks_number{4};
int results[checks_number]{};
int results[checks_number]{41, 42, 43, 44};
{
sycl::buffer<int> buf(data, sycl::range<1>(n));
sycl::buffer<int> results_buf(results, sycl::range<1>(checks_number));
sycl::queue q;
if (!core_sg_supported(q.get_device())) {
std::cout << "Skipping test" << std::endl;
return 0;
}
sycl::nd_range<1> NDR(sycl::range<1>{n}, sycl::range<1>{2});
q.submit([&](sycl::handler &cgh) {
sycl::accessor<int, 1, sycl::access::mode::write,
Expand Down
6 changes: 1 addition & 5 deletions sycl/test/e2e_test_requirements/no-unsupported-without-info.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -54,7 +54,7 @@
// tests to match the required format and in that case you should just update
// (i.e. reduce) the number and the list below.
//
// NUMBER-OF-UNSUPPORTED-WITHOUT-INFO: 280
// NUMBER-OF-UNSUPPORTED-WITHOUT-INFO: 276
//
// List of improperly UNSUPPORTED tests.
// Remove the CHECK once the test has been properly UNSUPPORTED.
Expand Down Expand Up @@ -86,10 +86,6 @@
// CHECK-NEXT: Assert/check_resource_leak.cpp
// CHECK-NEXT: Basic/buffer/buffer_create.cpp
// CHECK-NEXT: Basic/build_log.cpp
// CHECK-NEXT: Basic/free_function_queries/free_function_queries.cpp
// CHECK-NEXT: Basic/free_function_queries/free_function_queries.cpp
// CHECK-NEXT: Basic/free_function_queries/free_function_queries_sub_group.cpp
// CHECK-NEXT: Basic/free_function_queries/free_function_queries_sub_group.cpp
// CHECK-NEXT: Basic/gpu_max_wgs_error.cpp
// CHECK-NEXT: Basic/group_async_copy.cpp
// CHECK-NEXT: Basic/host-task-dependency.cpp
Expand Down
Loading