From ed49165d7b2a3016a2bc7933dc988fc7c2aacbd3 Mon Sep 17 00:00:00 2001
From: KornevNikita <nikita.kornev@intel.com>
Date: Wed, 23 Nov 2022 07:23:38 -0800
Subject: [PATCH 1/3] [SYCL] Test sycl::is_compatible()

---
 SYCL/OptionalKernelFeatures/is_compatible.cpp | 34 +++++++++++++++++++
 1 file changed, 34 insertions(+)
 create mode 100644 SYCL/OptionalKernelFeatures/is_compatible.cpp

diff --git a/SYCL/OptionalKernelFeatures/is_compatible.cpp b/SYCL/OptionalKernelFeatures/is_compatible.cpp
new file mode 100644
index 0000000000..47e59d8426
--- /dev/null
+++ b/SYCL/OptionalKernelFeatures/is_compatible.cpp
@@ -0,0 +1,34 @@
+// requires: cpu
+// RUN: %clangxx -fsycl -O0 %s -o %t.out
+// RUN: %CPU_RUN_PLACEHOLDER %t.out
+
+#include <sycl/sycl.hpp>
+
+class Kernel1;
+class Kernel2;
+
+[[sycl::device_has(sycl::aspect::cpu)]] void foo() {};
+[[sycl::device_has(sycl::aspect::gpu)]] void bar() {};
+
+int main() {
+  bool Passed = false;
+  sycl::device dev(sycl::cpu_selector_v);
+
+  if (sycl::is_compatible<Kernel1>(dev)) {
+    sycl::queue q{dev};
+    q.submit([&](sycl::handler &cgh) {
+        cgh.single_task<Kernel1>([=]() { foo(); });
+    }).wait_and_throw();
+    Passed = true;
+  }
+
+  if (sycl::is_compatible<Kernel2>(dev)) {
+    sycl::queue q{dev};
+    q.submit([&](sycl::handler &cgh) {
+        cgh.single_task<Kernel2>([=]() { bar(); });
+    }).wait_and_throw();
+    return 1;
+  }
+
+  return Passed ? 0 : 1;
+}
\ No newline at end of file

From 8d8f6d55c604101f11f47fd4a5ac7556c17d1b79 Mon Sep 17 00:00:00 2001
From: KornevNikita <nikita.kornev@intel.com>
Date: Wed, 23 Nov 2022 07:44:13 -0800
Subject: [PATCH 2/3] Add newline

---
 SYCL/OptionalKernelFeatures/is_compatible.cpp | 2 +-
 1 file changed, 1 insertion(+), 1 deletion(-)

diff --git a/SYCL/OptionalKernelFeatures/is_compatible.cpp b/SYCL/OptionalKernelFeatures/is_compatible.cpp
index 47e59d8426..9aab317d0e 100644
--- a/SYCL/OptionalKernelFeatures/is_compatible.cpp
+++ b/SYCL/OptionalKernelFeatures/is_compatible.cpp
@@ -31,4 +31,4 @@ int main() {
   }
 
   return Passed ? 0 : 1;
-}
\ No newline at end of file
+}

From 536c93ed161ac845c955b717f03f8b4e578def0b Mon Sep 17 00:00:00 2001
From: KornevNikita <nikita.kornev@intel.com>
Date: Fri, 25 Nov 2022 06:16:42 -0800
Subject: [PATCH 3/3] Adjustment

---
 SYCL/OptionalKernelFeatures/is_compatible.cpp | 54 +++++++++++--------
 1 file changed, 33 insertions(+), 21 deletions(-)

diff --git a/SYCL/OptionalKernelFeatures/is_compatible.cpp b/SYCL/OptionalKernelFeatures/is_compatible.cpp
index 9aab317d0e..d1f17326b9 100644
--- a/SYCL/OptionalKernelFeatures/is_compatible.cpp
+++ b/SYCL/OptionalKernelFeatures/is_compatible.cpp
@@ -1,34 +1,46 @@
-// requires: cpu
+// requires: cpu, gpu, accelerator
 // RUN: %clangxx -fsycl -O0 %s -o %t.out
 // RUN: %CPU_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
+// RUN: %ACC_RUN_PLACEHOLDER %t.out
 
 #include <sycl/sycl.hpp>
 
-class Kernel1;
-class Kernel2;
+[[sycl::device_has(sycl::aspect::cpu)]] void foo(){};
+[[sycl::device_has(sycl::aspect::gpu)]] void bar(){};
+[[sycl::device_has(sycl::aspect::accelerator)]] void baz(){};
 
-[[sycl::device_has(sycl::aspect::cpu)]] void foo() {};
-[[sycl::device_has(sycl::aspect::gpu)]] void bar() {};
+class KernelCPU;
+class KernelGPU;
+class KernelACC;
 
 int main() {
-  bool Passed = false;
-  sycl::device dev(sycl::cpu_selector_v);
+  bool Compatible = true;
+  bool Called = false;
+  sycl::device Dev;
+  sycl::queue Q(Dev);
 
-  if (sycl::is_compatible<Kernel1>(dev)) {
-    sycl::queue q{dev};
-    q.submit([&](sycl::handler &cgh) {
-        cgh.single_task<Kernel1>([=]() { foo(); });
-    }).wait_and_throw();
-    Passed = true;
+  if (sycl::is_compatible<KernelCPU>(Dev)) {
+    Q.submit(
+        [&](sycl::handler &h) { h.single_task<KernelCPU>([=]() { foo(); }); });
+    Q.wait();
+    Compatible &= Dev.is_cpu();
+    Called = true;
   }
-
-  if (sycl::is_compatible<Kernel2>(dev)) {
-    sycl::queue q{dev};
-    q.submit([&](sycl::handler &cgh) {
-        cgh.single_task<Kernel2>([=]() { bar(); });
-    }).wait_and_throw();
-    return 1;
+  if (sycl::is_compatible<KernelGPU>(Dev)) {
+    Q.submit(
+        [&](sycl::handler &h) { h.single_task<KernelGPU>([=]() { bar(); }); });
+    Q.wait();
+    Compatible &= Dev.is_gpu();
+    Called = true;
+  }
+  if (sycl::is_compatible<KernelACC>(Dev)) {
+    Q.submit(
+        [&](sycl::handler &h) { h.single_task<KernelACC>([=]() { baz(); }); });
+    Q.wait();
+    Compatible &= Dev.is_accelerator();
+    Called = true;
   }
 
-  return Passed ? 0 : 1;
+  return (Compatible && Called) ? 0 : 1;
 }