[clang] [clang][SYCL] Add sycl_external attribute and restrict emitting device code (PR #140282)

via cfe-commits cfe-commits at lists.llvm.org
Mon Jun 16 10:49:57 PDT 2025


================
@@ -0,0 +1,38 @@
+// RUN: %clang_cc1 -fsycl-is-device -fsyntax-only -verify %s
+
+// Semantic tests for sycl_external attribute
+
+[[clang::sycl_external(3)]] // expected-error {{'sycl_external' attribute takes no arguments}}
+void bar() {}
+
+[[clang::sycl_external]] // expected-error {{'sycl_external' can only be applied to functions with external linkage}}
+static void func1() {}
+
+namespace {
+  [[clang::sycl_external]] // expected-error {{'sycl_external' can only be applied to functions with external linkage}}
+  void func2() {}
+
+  struct UnnX {};
+}
+
+[[clang::sycl_external]] // expected-error {{'sycl_external' can only be applied to functions with external linkage}}
+  void func4(UnnX) {}
+
+class A {
+  [[clang::sycl_external]]
+  A() {}
+
+  [[clang::sycl_external]] void func3() {}
+};
+
+class B {
+public:
+  [[clang::sycl_external]] virtual void foo() {}
+
+  [[clang::sycl_external]] virtual void bar() = 0;
+};
+
+[[clang::sycl_external]] int *func0() { return nullptr; }
+
+[[clang::sycl_external]] void func2(int *) {}
+
----------------
schittir wrote:

I created a tracker to account for the following missing diagnostics that need to be implemented to follow spec 2020 for this attribute. 

- When compiling for a SYCL device target that does not support the generic address space, the function shall not specify a raw pointer or reference type as the return type or as a parameter type.
- The function cannot call sycl::group::parallel_for_work_item.
- The function cannot be called from a parallel_for_work_group scope.

https://github.com/llvm/llvm-project/pull/140282


More information about the cfe-commits mailing list