[SYCL] Mark tests that require size 32 subgroups with "REQUIRES: sg-32" (#11184)

Many AMD devices don't support size 32 subgroups. This patch skips tests
for such devices. See also #11104.

Author: hdelan

sycl/test-e2e/NonUniformGroups/fixed_size_group.cpp

@@ -2,6 +2,7 @@
 // RUN: %{run} %t.out
 
 // REQUIRES: gpu
+// REQUIRES: sg-32
 
 #include <sycl/sycl.hpp>
 #include <vector>
@@ -12,13 +13,6 @@ template <size_t PartitionSize> class TestKernel;
 template <size_t PartitionSize> void test() {
   sycl::queue Q;
 
-  auto SGSizes = Q.get_device().get_info<sycl::info::device::sub_group_sizes>();
-  if (std::find(SGSizes.begin(), SGSizes.end(), 32) == SGSizes.end()) {
-    std::cout << "Test skipped due to missing support for sub-group size 32."
-              << std::endl;
-    return;
-  }
-
   sycl::buffer<bool, 1> MatchBuf{sycl::range{32}};
   sycl::buffer<bool, 1> LeaderBuf{sycl::range{32}};
 

sycl/test-e2e/NonUniformGroups/fixed_size_group_algorithms.cpp

@@ -2,6 +2,7 @@
 // RUN: %{run} %t.out
 //
 // REQUIRES: gpu
+// REQUIRES: sg-32
 // UNSUPPORTED: hip
 
 #include <sycl/sycl.hpp>
@@ -14,12 +15,6 @@ template <size_t PartitionSize> void test() {
   sycl::queue Q;
 
   constexpr uint32_t SGSize = 32;
-  auto SGSizes = Q.get_device().get_info<sycl::info::device::sub_group_sizes>();
-  if (std::find(SGSizes.begin(), SGSizes.end(), SGSize) == SGSizes.end()) {
-    std::cout << "Test skipped due to missing support for sub-group size 32."
-              << std::endl;
-    return;
-  }
 
   sycl::buffer<size_t, 1> TmpBuf{sycl::range{SGSize}};
   sycl::buffer<bool, 1> BarrierBuf{sycl::range{SGSize}};

sycl/test-e2e/NonUniformGroups/tangle_group_algorithms.cpp

@@ -2,6 +2,7 @@
 // RUN: %{run} %t.out
 //
 // REQUIRES: gpu
+// REQUIRES: sg-32
 // UNSUPPORTED: cuda || hip || windows
 // Tangle groups exhibit unpredictable behavior on Windows.
 // The test is disabled while we investigate the root cause.
@@ -17,13 +18,6 @@ constexpr uint32_t SGSize = 32;
 int main() {
   sycl::queue Q;
 
-  auto SGSizes = Q.get_device().get_info<sycl::info::device::sub_group_sizes>();
-  if (std::find(SGSizes.begin(), SGSizes.end(), SGSize) == SGSizes.end()) {
-    std::cout << "Test skipped due to missing support for sub-group size 32."
-              << std::endl;
-    return 0;
-  }
-
   sycl::buffer<size_t, 1> TmpBuf{sycl::range{SGSize}};
   sycl::buffer<bool, 1> BarrierBuf{sycl::range{SGSize}};
   sycl::buffer<bool, 1> BroadcastBuf{sycl::range{SGSize}};

sycl/test-e2e/SubGroupMask/Basic.cpp

@@ -1,6 +1,7 @@
 // RUN: %{build} -o %t.out
 
 // REQUIRES: gpu
+// REQUIRES: sg-32
 
 // GroupNonUniformBallot capability is supported on Intel GPU only
 // RUN: %{run} %t.out
@@ -24,14 +25,6 @@ int main() {
 #ifdef SYCL_EXT_ONEAPI_SUB_GROUP_MASK
   queue Queue;
 
-  auto sgsizes =
-      Queue.get_device().get_info<sycl::info::device::sub_group_sizes>();
-  if (std::find(sgsizes.begin(), sgsizes.end(), 32) == sgsizes.end()) {
-    std::cout << "test skipped due to missing support for sub-group size 32."
-              << std::endl;
-    return 0;
-  }
-
   try {
     nd_range<1> NdRange(global_size, local_size);
     int Res = 0;

sycl/test-e2e/syclcompat/util/util_logical_group.cpp

@@ -30,6 +30,7 @@
 //
 // ===----------------------------------------------------------------------===//
 
+// REQUIRES: sg-32
 // RUN: %clangxx -fsycl -fsycl-targets=%{sycl_triple} %s -o %t.out
 // RUN: %{run} %t.out
 

sycl/test-e2e/syclcompat/util/util_permute_sub_group_by_xor.cpp

@@ -30,6 +30,7 @@
 //
 // ===----------------------------------------------------------------------===//
 
+// REQUIRES: sg-32
 // RUN: %clangxx -fsycl -fsycl-targets=%{sycl_triple} %s -o %t.out
 // RUN: %{run} %t.out
 

sycl/test-e2e/syclcompat/util/util_select_from_sub_group.cpp

@@ -30,6 +30,7 @@
 //
 // ===----------------------------------------------------------------------===//
 
+// REQUIRES: sg-32
 // RUN: %clangxx -fsycl -fsycl-targets=%{sycl_triple} %s -o %t.out
 // RUN: %{run} %t.out
 

sycl/test-e2e/syclcompat/util/util_shift_sub_group_left.cpp

@@ -30,6 +30,7 @@
 //
 // ===----------------------------------------------------------------------===//
 
+// REQUIRES: sg-32
 // RUN: %clangxx -fsycl -fsycl-targets=%{sycl_triple} %s -o %t.out
 // RUN: %{run} %t.out
 

sycl/test-e2e/syclcompat/util/util_shift_sub_group_right.cpp

@@ -30,6 +30,7 @@
 //
 // ===----------------------------------------------------------------------===//
 
+// REQUIRES: sg-32
 // RUN: %clangxx -fsycl -fsycl-targets=%{sycl_triple} %s -o %t.out
 // RUN: %{run} %t.out