Don't use 0 for named_sub_group_size value

Author: jzc

sycl/include/sycl/ext/oneapi/experimental/named_sub_group_sizes.hpp

@@ -15,8 +15,8 @@ inline namespace _V1 {
 namespace ext::oneapi::experimental {
 
 struct named_sub_group_size {
-  static constexpr uint32_t primary = 0;
-  static constexpr uint32_t automatic = -1;
+  static constexpr uint32_t primary = -1;
+  static constexpr uint32_t automatic = -2;
 };
 
 inline constexpr sub_group_size_key::value_t<named_sub_group_size::primary>

sycl/test/extensions/properties/properties_kernel_named_sub_group_size.cpp

@@ -21,4 +21,4 @@ int main() {
   q.parallel_for<class Kernel2>(ndr, P2, [=](auto id) {});
 }
 
-// CHECK: ![[SGSizeAttr]] = !{i32 0}
+// CHECK: ![[SGSizeAttr]] = !{i32 -1}

sycl/test/extensions/properties/properties_kernel_negative.cpp

@@ -301,6 +301,11 @@ void check_sub_group_size() {
   // expected-error@+1 {{too few template arguments for variable template 'sub_group_size'}}
   auto WGSize0 = sycl::ext::oneapi::experimental::sub_group_size<>;
 
+  // expected-error-re@sycl/ext/oneapi/kernel_properties/properties.hpp:* {{static assertion failed due to requirement {{.+}}: sub_group_size_key property must contain a non-zero value.}}
+  // expected-error@sycl/ext/oneapi/kernel_properties/properties.hpp:* {{constexpr variable 'sub_group_size<0>' must be initialized by a constant expression}}
+  // expected-note@+1 {{in instantiation of variable template specialization 'sycl::ext::oneapi::experimental::sub_group_size<0>' requested here}}
+  auto WGSize1 = sycl::ext::oneapi::experimental::sub_group_size<0>;
+
   sycl::queue Q;
 
   // expected-error-re@sycl/ext/oneapi/properties/properties.hpp:* {{static assertion failed due to requirement {{.+}}: Failed to merge property lists due to conflicting properties.}}