Isolate fixed sub-group kernel for CTS.

Author: isaacault

test/conformance/device_code/CMakeLists.txt

@@ -159,6 +159,7 @@ add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/saxpy.cpp)
 add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/saxpy_usm.cpp)
 add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/indexers_usm.cpp)
 add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/build_failure.cpp)
+add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/fixed_sg_size.cpp)
 add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/fixed_wg_size.cpp)
 add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/max_wg_size.cpp)
 add_device_binary(${CMAKE_CURRENT_SOURCE_DIR}/sequence.cpp)

test/conformance/device_code/fixed_sg_size.cpp

@@ -0,0 +1,27 @@
+// Copyright (C) 2024 Intel Corporation
+// Part of the Unified-Runtime Project, under the Apache License v2.0 with LLVM Exceptions.
+// See LICENSE.TXT
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+
+#include <sycl/sycl.hpp>
+
+struct KernelFunctor {
+    void operator()(sycl::nd_item<3>) const {}
+    void operator()(sycl::item<3>) const {}
+
+    auto get(sycl::ext::oneapi::experimental::properties_tag) {
+        return sycl::ext::oneapi::experimental::properties{
+            sycl::ext::oneapi::experimental::sub_group_size<8>};
+    }
+};
+
+int main() {
+    sycl::queue myQueue;
+    myQueue.submit([&](sycl::handler &cgh) {
+        cgh.parallel_for<class FixedSgSize>(sycl::range<3>(8, 8, 8),
+                                            KernelFunctor{});
+    });
+
+    myQueue.wait();
+    return 0;
+}

test/conformance/device_code/fixed_wg_size.cpp

@@ -11,8 +11,7 @@ struct KernelFunctor {
 
     auto get(sycl::ext::oneapi::experimental::properties_tag) {
         return sycl::ext::oneapi::experimental::properties{
-            sycl::ext::oneapi::experimental::work_group_size<8, 4, 2>,
-            sycl::ext::oneapi::experimental::sub_group_size<8>};
+            sycl::ext::oneapi::experimental::work_group_size<8, 4, 2>};
     }
 };
 

test/conformance/kernel/kernel_adapter_level_zero.match

@@ -1,8 +1,6 @@
-# Match tests that use fixed_wg_size.cpp as it fails to compile on some 
+# Match tests that use fixed_sg_size.cpp as it fails to compile on some 
 # hardware.
-urKernelGetGroupInfoMaxWorkGroupSizeTest.CompileMaxWorkGroupSize/*
-urKernelGetGroupInfoMaxWorkGroupSizeTest.CompileMaxLinearWorkGroupSize/*
-urKernelGetSubGroupInfoFixedSubGroupSizeTest.CompileNumSubGroups/*
+{{OPT}}urKernelGetSubGroupInfoFixedSubGroupSizeTest.CompileNumSubGroups/*
 
 urKernelSetExecInfoTest.SuccessIndirectAccess/*
 urKernelSetExecInfoUSMPointersTest.SuccessHost/*

test/conformance/kernel/kernel_adapter_level_zero_v2.match

@@ -1,8 +1,6 @@
-# Match tests that use fixed_wg_size.cpp as it fails to compile on some 
+# Match tests that use fixed_sg_size.cpp as it fails to compile on some 
 # hardware.
-urKernelGetGroupInfoMaxWorkGroupSizeTest.CompileMaxWorkGroupSize/*
-urKernelGetGroupInfoMaxWorkGroupSizeTest.CompileMaxLinearWorkGroupSize/*
-urKernelGetSubGroupInfoFixedSubGroupSizeTest.CompileNumSubGroups/*
+{{OPT}}urKernelGetSubGroupInfoFixedSubGroupSizeTest.CompileNumSubGroups/*
 
 urKernelSetExecInfoTest.SuccessIndirectAccess/*
 urKernelSetExecInfoUSMPointersTest.SuccessHost/*

test/conformance/kernel/kernel_adapter_opencl.match

@@ -1,5 +1,8 @@
 # Match tests that use fixed_wg_size.cpp as it fails to compile on some 
 # hardware.
-urKernelGetGroupInfoMaxWorkGroupSizeTest.CompileMaxWorkGroupSize/*
-urKernelGetGroupInfoMaxWorkGroupSizeTest.CompileMaxLinearWorkGroupSize/*
-urKernelGetSubGroupInfoFixedSubGroupSizeTest.CompileNumSubGroups/*
+{{OPT}}urKernelGetGroupInfoMaxWorkGroupSizeTest.CompileMaxWorkGroupSize/*
+{{OPT}}urKernelGetGroupInfoMaxWorkGroupSizeTest.CompileMaxLinearWorkGroupSize/*
+
+# Match tests that use fixed_sg_size.cpp as it fails to compile on some 
+# hardware.
+{{OPT}}urKernelGetSubGroupInfoFixedSubGroupSizeTest.CompileNumSubGroups/*

test/conformance/kernel/urKernelGetSubGroupInfo.cpp

@@ -8,11 +8,11 @@
 
 struct urKernelGetSubGroupInfoFixedSubGroupSizeTest : uur::urKernelTest {
     void SetUp() override {
-        program_name = "fixed_wg_size";
+        program_name = "fixed_sg_size";
         UUR_RETURN_ON_FATAL_FAILURE(urKernelTest::SetUp());
     }
 
-    // This value correlates to sub_group_size<8> in fixed_wg_size.cpp.
+    // This value correlates to sub_group_size<8> in fixed_sg_size.cpp.
     uint32_t num_sub_groups{8};
 };
 UUR_INSTANTIATE_DEVICE_TEST_SUITE_P(