From 18ecc5a43daf5f6e1cb4a11dc3146fcf19e9f288 Mon Sep 17 00:00:00 2001
From: "Larsen, Steffen" <steffen.larsen@intel.com>
Date: Fri, 14 Nov 2025 07:18:04 -0800
Subject: [PATCH 1/2] [SYCL][clang] Add attr/props warning case for multiple
 attributes

This commit adds new cases to the existing tests for checking the
warning issued when there may be potential conflicts between SYCL
attributes and compile-time properties. These tests check that multiple
warning diagnostics are issued if there are multiple attributes on the
same kernel, each specifying the potentially conflicting attribute.

Signed-off-by: Larsen, Steffen <steffen.larsen@intel.com>
---
 ...tr-add-ir-attributes-function-conflict.cpp | 20 +++++++++++++++++
 .../properties_kernel_negative_device.cpp     | 22 +++++++++++++++++++
 2 files changed, 42 insertions(+)

diff --git a/clang/test/SemaSYCL/attr-add-ir-attributes-function-conflict.cpp b/clang/test/SemaSYCL/attr-add-ir-attributes-function-conflict.cpp
index 5fec0b4961618..7f683c70f8f90 100644
--- a/clang/test/SemaSYCL/attr-add-ir-attributes-function-conflict.cpp
+++ b/clang/test/SemaSYCL/attr-add-ir-attributes-function-conflict.cpp
@@ -112,6 +112,11 @@ int main() {
   NonemptyWrapper.kernel_single_task<class NEK9>([]() [[sycl::device_has(sycl::aspect::cpu)]] {});
   // expected-warning@+1 {{kernel has both attribute 'sycl::device_has' and kernel properties; conflicting properties are ignored}}
   NonemptyWrapper.kernel_single_task<class NEK10>([]() [[sycl::device_has(sycl::aspect::cpu, sycl::aspect::gpu)]] {});
+  // expected-warning@+4 {{kernel has both attribute 'sycl::reqd_work_group_size' and kernel properties; conflicting properties are ignored}}
+  // expected-warning@+3 {{kernel has both attribute 'sycl::work_group_size_hint' and kernel properties; conflicting properties are ignored}}
+  // expected-warning@+2 {{kernel has both attribute 'sycl::reqd_sub_group_size' and kernel properties; conflicting properties are ignored}}
+  // expected-warning@+1 {{kernel has both attribute 'sycl::device_has' and kernel properties; conflicting properties are ignored}}
+  NonemptyWrapper.kernel_single_task<class NEK11>([]() [[sycl::reqd_work_group_size(1)]] [[sycl::work_group_size_hint(1)]] [[sycl::reqd_sub_group_size(1)]] [[sycl::device_has(sycl::aspect::cpu, sycl::aspect::gpu)]] {});
 
   // expected-warning@+1 {{kernel has both attribute 'sycl::reqd_work_group_size' and kernel properties; conflicting properties are ignored}}
   NonemptyWrapperWithImplicit.kernel_single_task<class NEKWI1>([]() [[sycl::reqd_work_group_size(1)]] {});
@@ -133,6 +138,11 @@ int main() {
   NonemptyWrapperWithImplicit.kernel_single_task<class NEKWI9>([]() [[sycl::device_has(sycl::aspect::cpu)]] {});
   // expected-warning@+1 {{kernel has both attribute 'sycl::device_has' and kernel properties; conflicting properties are ignored}}
   NonemptyWrapperWithImplicit.kernel_single_task<class NEKWI10>([]() [[sycl::device_has(sycl::aspect::cpu, sycl::aspect::gpu)]] {});
+  // expected-warning@+4 {{kernel has both attribute 'sycl::reqd_work_group_size' and kernel properties; conflicting properties are ignored}}
+  // expected-warning@+3 {{kernel has both attribute 'sycl::work_group_size_hint' and kernel properties; conflicting properties are ignored}}
+  // expected-warning@+2 {{kernel has both attribute 'sycl::reqd_sub_group_size' and kernel properties; conflicting properties are ignored}}
+  // expected-warning@+1 {{kernel has both attribute 'sycl::device_has' and kernel properties; conflicting properties are ignored}}
+  NonemptyWrapperWithImplicit.kernel_single_task<class NEKWI11>([]() [[sycl::reqd_work_group_size(1)]] [[sycl::work_group_size_hint(1)]] [[sycl::reqd_sub_group_size(1)]] [[sycl::device_has(sycl::aspect::cpu, sycl::aspect::gpu)]] {});
 
   // expected-warning@+1 {{kernel has both attribute 'sycl::reqd_work_group_size' and kernel properties; conflicting properties are ignored}}
   NonemptyWrapperWithFilter.kernel_single_task<class NEKWF1>([]() [[sycl::reqd_work_group_size(1)]] {});
@@ -154,6 +164,11 @@ int main() {
   NonemptyWrapperWithFilter.kernel_single_task<class NEKWF9>([]() [[sycl::device_has(sycl::aspect::cpu)]] {});
   // expected-warning@+1 {{kernel has both attribute 'sycl::device_has' and kernel properties; conflicting properties are ignored}}
   NonemptyWrapperWithFilter.kernel_single_task<class NEKWF10>([]() [[sycl::device_has(sycl::aspect::cpu, sycl::aspect::gpu)]] {});
+  // expected-warning@+4 {{kernel has both attribute 'sycl::reqd_work_group_size' and kernel properties; conflicting properties are ignored}}
+  // expected-warning@+3 {{kernel has both attribute 'sycl::work_group_size_hint' and kernel properties; conflicting properties are ignored}}
+  // expected-warning@+2 {{kernel has both attribute 'sycl::reqd_sub_group_size' and kernel properties; conflicting properties are ignored}}
+  // expected-warning@+1 {{kernel has both attribute 'sycl::device_has' and kernel properties; conflicting properties are ignored}}
+  NonemptyWrapperWithFilter.kernel_single_task<class NEKWF11>([]() [[sycl::reqd_work_group_size(1)]] [[sycl::work_group_size_hint(1)]] [[sycl::reqd_sub_group_size(1)]] [[sycl::device_has(sycl::aspect::cpu, sycl::aspect::gpu)]] {});
 
   // expected-warning@+1 {{kernel has both attribute 'sycl::reqd_work_group_size' and kernel properties; conflicting properties are ignored}}
   NonemptyWrapperWithImplicitAndFilter.kernel_single_task<class NEKWIF1>([]() [[sycl::reqd_work_group_size(1)]] {});
@@ -175,4 +190,9 @@ int main() {
   NonemptyWrapperWithImplicitAndFilter.kernel_single_task<class NEKWIF9>([]() [[sycl::device_has(sycl::aspect::cpu)]] {});
   // expected-warning@+1 {{kernel has both attribute 'sycl::device_has' and kernel properties; conflicting properties are ignored}}
   NonemptyWrapperWithImplicitAndFilter.kernel_single_task<class NEKWIF10>([]() [[sycl::device_has(sycl::aspect::cpu, sycl::aspect::gpu)]] {});
+  // expected-warning@+4 {{kernel has both attribute 'sycl::reqd_work_group_size' and kernel properties; conflicting properties are ignored}}
+  // expected-warning@+3 {{kernel has both attribute 'sycl::work_group_size_hint' and kernel properties; conflicting properties are ignored}}
+  // expected-warning@+2 {{kernel has both attribute 'sycl::reqd_sub_group_size' and kernel properties; conflicting properties are ignored}}
+  // expected-warning@+1 {{kernel has both attribute 'sycl::device_has' and kernel properties; conflicting properties are ignored}}
+  NonemptyWrapperWithImplicitAndFilter.kernel_single_task<class NEKWIF11>([]() [[sycl::reqd_work_group_size(1)]] [[sycl::work_group_size_hint(1)]] [[sycl::reqd_sub_group_size(1)]] [[sycl::device_has(sycl::aspect::cpu, sycl::aspect::gpu)]] {});
 }
diff --git a/sycl/test/extensions/properties/properties_kernel_negative_device.cpp b/sycl/test/extensions/properties/properties_kernel_negative_device.cpp
index 44718aa5f3125..c5576667216f3 100644
--- a/sycl/test/extensions/properties/properties_kernel_negative_device.cpp
+++ b/sycl/test/extensions/properties/properties_kernel_negative_device.cpp
@@ -38,6 +38,21 @@ template <sycl::aspect Aspect> struct KernelFunctorWithDeviceHasWithAttr {
   }
 };
 
+template <sycl::aspect Aspect> struct KernelFunctorWithAllAttrs {
+  // expected-warning@+4 {{kernel has both attribute 'sycl::reqd_work_group_size' and kernel properties; conflicting properties are ignored}}
+  // expected-warning@+3 {{kernel has both attribute 'sycl::work_group_size_hint' and kernel properties; conflicting properties are ignored}}
+  // expected-warning@+2 {{kernel has both attribute 'sycl::reqd_sub_group_size' and kernel properties; conflicting properties are ignored}}
+  // expected-warning@+1 {{kernel has both attribute 'sycl::device_has' and kernel properties; conflicting properties are ignored}}
+  void operator() [[sycl::reqd_work_group_size(
+      32)]] [[sycl::
+                  work_group_size_hint(32)]] [[sycl::reqd_sub_group_size(
+      32)]] [[sycl::device_has(sycl::aspect::cpu)]] () const {}
+  auto get(sycl::ext::oneapi::experimental::properties_tag) const {
+    return sycl::ext::oneapi::experimental::properties{
+        sycl::ext::oneapi::experimental::device_has<Aspect>};
+  }
+};
+
 void check_work_group_size() {
   sycl::queue Q;
 
@@ -62,10 +77,17 @@ void check_device_has() {
   Q.single_task(KernelFunctorWithDeviceHasWithAttr<sycl::aspect::cpu>{});
 }
 
+void check_all() {
+  sycl::queue Q;
+
+  Q.single_task(KernelFunctorWithAllAttrs<sycl::aspect::cpu>{});
+}
+
 int main() {
   check_work_group_size();
   check_work_group_size_hint();
   check_sub_group_size();
   check_device_has();
+  check_all();
   return 0;
 }

From 677b98ca7b3869ea61468ec9b57e23db3e11b7cb Mon Sep 17 00:00:00 2001
From: "Larsen, Steffen" <steffen.larsen@intel.com>
Date: Tue, 18 Nov 2025 23:33:35 -0800
Subject: [PATCH 2/2] Fix expected lines after formatting

Signed-off-by: Larsen, Steffen <steffen.larsen@intel.com>
---
 .../properties/properties_kernel_negative_device.cpp        | 6 +++---
 1 file changed, 3 insertions(+), 3 deletions(-)

diff --git a/sycl/test/extensions/properties/properties_kernel_negative_device.cpp b/sycl/test/extensions/properties/properties_kernel_negative_device.cpp
index c5576667216f3..2e34cc6a39a55 100644
--- a/sycl/test/extensions/properties/properties_kernel_negative_device.cpp
+++ b/sycl/test/extensions/properties/properties_kernel_negative_device.cpp
@@ -40,9 +40,9 @@ template <sycl::aspect Aspect> struct KernelFunctorWithDeviceHasWithAttr {
 
 template <sycl::aspect Aspect> struct KernelFunctorWithAllAttrs {
   // expected-warning@+4 {{kernel has both attribute 'sycl::reqd_work_group_size' and kernel properties; conflicting properties are ignored}}
-  // expected-warning@+3 {{kernel has both attribute 'sycl::work_group_size_hint' and kernel properties; conflicting properties are ignored}}
-  // expected-warning@+2 {{kernel has both attribute 'sycl::reqd_sub_group_size' and kernel properties; conflicting properties are ignored}}
-  // expected-warning@+1 {{kernel has both attribute 'sycl::device_has' and kernel properties; conflicting properties are ignored}}
+  // expected-warning@+4 {{kernel has both attribute 'sycl::work_group_size_hint' and kernel properties; conflicting properties are ignored}}
+  // expected-warning@+4 {{kernel has both attribute 'sycl::reqd_sub_group_size' and kernel properties; conflicting properties are ignored}}
+  // expected-warning@+4 {{kernel has both attribute 'sycl::device_has' and kernel properties; conflicting properties are ignored}}
   void operator() [[sycl::reqd_work_group_size(
       32)]] [[sycl::
                   work_group_size_hint(32)]] [[sycl::reqd_sub_group_size(