Add missing code to enforce subgroup size in subgroup test.

Author: aarongreig

test/conformance/device_code/subgroup.cpp

@@ -11,6 +11,11 @@ struct KernelFunctor {
     KernelFunctor(sycl::accessor<size_t, 1, sycl::access_mode::write> Acc)
         : Acc(Acc) {}
 
+    auto get(sycl::ext::oneapi::experimental::properties_tag) {
+        return sycl::ext::oneapi::experimental::properties{
+            sycl::ext::oneapi::experimental::sub_group_size<8>};
+    }
+
     void operator()(sycl::nd_item<1> NdItem) const {
         auto SG = NdItem.get_sub_group();
         if (NdItem.get_global_linear_id() == 0) {

test/conformance/enqueue/enqueue_adapter_level_zero_v2.match

@@ -2,7 +2,6 @@
 urEnqueueDeviceGetGlobalVariableReadTest.Success/Intel_R__oneAPI_Unified_Runtime_over_Level_Zero___{{.*}}_
 urEnqueueKernelLaunchTest.InvalidKernelArgs/Intel_R__oneAPI_Unified_Runtime_over_Level_Zero___{{.*}}_
 urEnqueueKernelLaunchKernelWgSizeTest.Success/Intel_R__oneAPI_Unified_Runtime_over_Level_Zero___{{.*}}_
-urEnqueueKernelLaunchKernelSubGroupTest.Success/Intel_R__oneAPI_Unified_Runtime_over_Level_Zero___{{.*}}_
 urEnqueueKernelLaunchWithVirtualMemory.Success/Intel_R__oneAPI_Unified_Runtime_over_Level_Zero___{{.*}}_
 urEnqueueKernelLaunchUSMLinkedList.Success/Intel_R__oneAPI_Unified_Runtime_over_Level_Zero___{{.*}}___UsePoolEnabled
 urEnqueueKernelLaunchUSMLinkedList.Success/Intel_R__oneAPI_Unified_Runtime_over_Level_Zero___{{.*}}___UsePoolDisabled

test/conformance/enqueue/enqueue_adapter_opencl.match

@@ -1,5 +1,4 @@
 {{NONDETERMINISTIC}}
 {{OPT}}urEnqueueDeviceGetGlobalVariableReadTest.Success/Intel_R__OpenCL___{{.*}}_
 urEnqueueKernelLaunchKernelWgSizeTest.Success/Intel_R__OpenCL___{{.*}}_
-urEnqueueKernelLaunchKernelSubGroupTest.Success/Intel_R__OpenCL___{{.*}}_
 {{OPT}}urEnqueueKernelLaunchUSMLinkedList.Success/Intel_R__OpenCL___{{.*}}_UsePoolEnabled

test/conformance/enqueue/urEnqueueKernelLaunch.cpp

@@ -180,6 +180,8 @@ TEST_P(urEnqueueKernelLaunchKernelSubGroupTest, Success) {
         queue, kernel, n_dimensions, global_offset.data(), global_size.data(),
         nullptr, 0, nullptr, nullptr));
     ASSERT_SUCCESS(urQueueFinish(queue));
+    // We specify this subgroup size in the kernel source, and then the kernel
+    // queries for its subgroup size at runtime and writes it to the buffer.
     ValidateBuffer<size_t>(buffer, sizeof(size_t), 8);
 }