[clang] [libclc] [llvm] [openmp] [Clang] `__attribute__((assume))` refactor (PR #84934)

via cfe-commits cfe-commits at lists.llvm.org
Tue Mar 12 08:41:59 PDT 2024


https://github.com/Sirraide created https://github.com/llvm/llvm-project/pull/84934

@AaronBallman, @erichkeane This is a followup to #81014 and #84582. For anyone not familiar with what this is about: Clang provides `__attribute__((assume))` and `[[clang::assume]]` as nonstandard spellings for the `[[omp::assume]]` attribute; this results in a potentially very confusing name clash with C++23’s `[[assume]]` attribute. This pr is about deciding what to do about this.

Just to see how much would be affected by this, I’ve gone ahead and removed every last mention of `__attribute__((assume))` and replaced it with `[[omp::assume]]`. Here are a few things I have observed / changes I’ve implemented:
- `[[omp::assume]]` can no longer be spelt `__attribute__((assume))` or `[[clang::assume]]`.
- `__attribute__((assume))` and `[[clang::assume]]` are now alternative spellings for C++23’s `[[assume]]`; this shouldn’t cause any problems due to differences in appertainment: the OMP attribute only appertains to functions, and C++23’s only to empty statements, so any code using `__attribute__((assume))` in that position with Clang would not have compiled anyway. This should help improve GCC compatibility.
- The OpenMP specification only supports `assume` as an attribute in C++11 and later. However, libclc appears to be using `__attribute__((assume))` internally, specifically, in one header that defines a macro that is then used throughout the codebase. I’m not familiar with libclc or OpenCL, so I’ve added `omp_assume` as an alternative spelling for the OpenMP attribute and replaced the attribute in the header in question with `__attribute__((__omp_assume__))`. 

  If it turns out that using `[[]]`-style attributes in libclc is fine after all, I would suggest using that instead and not providing the `omp_assume` spelling in the first place. It should be noted that, without the `omp_assume` spelling, it would be impossible to use this attribute in C without running into `-pedantic` warnings; we could consider supporting `[[omp::assume]]` in C23, though.
- From what I can tell, no-one except libclc is actually using this attribute? At least on github, the only matches I’ve found for `__attribute__((assume("omp` are in LLVM and various forks thereof. Given that it’s not particularly widely used, I don’t think removal *without* deprecation would be that big of a deal in this case, though if you think that that’s a non-option, then I’d instead suggest deprecating it and removing it in a later version of Clang.

TODO:
- [ ] Discuss whether this is what we want to do, or decide on what to do instead.
- [ ] Figure out whether we can use `[[omp::assume]]` in libclc.
- [ ] If we decide to support `[[omp::assume]]` in C (or an alternative spelling thereof, in which case we should also update diagnostics that mention `[[omp::assume]]` to display that spelling instead if we’re compiling C, if possible), add some tests for that.
- [ ] Update the documentation for `__attribute__((assume))`.

>From 79ae39d195e8332c8fd154a5184247312554ddb1 Mon Sep 17 00:00:00 2001
From: Sirraide <aeternalmail at gmail.com>
Date: Tue, 12 Mar 2024 16:09:47 +0100
Subject: [PATCH] [Clang] __attribute__((assume)) refactor

---
 clang/include/clang/Basic/Attr.td             |  5 +-
 clang/include/clang/Basic/AttrDocs.td         |  4 +-
 .../clang/Basic/DiagnosticSemaKinds.td        |  3 -
 clang/lib/Parse/ParseDecl.cpp                 |  7 ++
 clang/lib/Sema/SemaStmtAttr.cpp               |  3 +-
 clang/test/CodeGen/assume_attr.c              | 58 --------------
 clang/test/CodeGenCXX/assume_attr.cpp         | 48 +++++------
 clang/test/OpenMP/assumes_codegen.cpp         | 80 +++++++++----------
 clang/test/OpenMP/assumes_print.cpp           |  6 +-
 clang/test/OpenMP/assumes_template_print.cpp  | 20 ++---
 ...rallel_in_multiple_target_state_machines.c |  8 +-
 ...remarks_parallel_in_target_state_machine.c |  4 +-
 clang/test/Sema/attr-assume.c                 | 14 ----
 clang/test/SemaCXX/cxx23-assume.cpp           |  4 +
 libclc/generic/include/clc/clcfunc.h          |  2 +-
 llvm/lib/Transforms/IPO/OpenMPOpt.cpp         |  4 +-
 .../OpenMP/custom_state_machines.ll           |  2 +-
 .../OpenMP/custom_state_machines_pre_lto.ll   |  2 +-
 .../OpenMP/custom_state_machines_remarks.ll   | 10 +--
 llvm/test/Transforms/OpenMP/spmdization.ll    |  4 +-
 .../Transforms/OpenMP/spmdization_guarding.ll |  4 +-
 .../Transforms/OpenMP/spmdization_remarks.ll  | 14 ++--
 openmp/docs/remarks/OMP121.rst                |  6 +-
 openmp/docs/remarks/OMP133.rst                |  6 +-
 openmp/docs/remarks/OptimizationRemarks.rst   |  4 +-
 25 files changed, 130 insertions(+), 192 deletions(-)
 delete mode 100644 clang/test/CodeGen/assume_attr.c
 delete mode 100644 clang/test/Sema/attr-assume.c

diff --git a/clang/include/clang/Basic/Attr.td b/clang/include/clang/Basic/Attr.td
index 080340669b60a0..39fccc720bc9cd 100644
--- a/clang/include/clang/Basic/Attr.td
+++ b/clang/include/clang/Basic/Attr.td
@@ -1581,10 +1581,11 @@ def Unlikely : StmtAttr {
 def : MutualExclusions<[Likely, Unlikely]>;
 
 def CXXAssume : StmtAttr {
-  let Spellings = [CXX11<"", "assume", 202207>];
+  let Spellings = [CXX11<"", "assume", 202207>, Clang<"assume">];
   let Subjects = SubjectList<[NullStmt], ErrorDiag, "empty statements">;
   let Args = [ExprArgument<"Assumption">];
   let Documentation = [CXXAssumeDocs];
+  let HasCustomParsing = 1;
 }
 
 def NoMerge : DeclOrStmtAttr {
@@ -4159,7 +4160,7 @@ def OMPDeclareVariant : InheritableAttr {
 }
 
 def OMPAssume : InheritableAttr {
-  let Spellings = [Clang<"assume">, CXX11<"omp", "assume">];
+  let Spellings = [CXX11<"omp", "assume">, Clang<"omp_assume">];
   let Subjects = SubjectList<[Function, ObjCMethod]>;
   let InheritEvenIfAlreadyPresent = 1;
   let Documentation = [OMPAssumeDocs];
diff --git a/clang/include/clang/Basic/AttrDocs.td b/clang/include/clang/Basic/AttrDocs.td
index 2c07cd09b0d5b7..855d57228c56fc 100644
--- a/clang/include/clang/Basic/AttrDocs.td
+++ b/clang/include/clang/Basic/AttrDocs.td
@@ -4661,7 +4661,7 @@ def OMPAssumeDocs : Documentation {
   let Category = DocCatFunction;
   let Heading = "assume";
   let Content = [{
-Clang supports the ``__attribute__((assume("assumption")))`` attribute to
+Clang supports the ``[[omp::assume("assumption")]]`` attribute to
 provide additional information to the optimizer. The string-literal, here
 "assumption", will be attached to the function declaration such that later
 analysis and optimization passes can assume the "assumption" to hold.
@@ -4673,7 +4673,7 @@ A function can have multiple assume attributes and they propagate from prior
 declarations to later definitions. Multiple assumptions are aggregated into a
 single comma separated string. Thus, one can provide multiple assumptions via
 a comma separated string, i.a.,
-``__attribute__((assume("assumption1,assumption2")))``.
+``[[omp::assume("assumption1,assumption2")]]``.
 
 While LLVM plugins might provide more assumption strings, the default LLVM
 optimization passes are aware of the following assumptions:
diff --git a/clang/include/clang/Basic/DiagnosticSemaKinds.td b/clang/include/clang/Basic/DiagnosticSemaKinds.td
index c54105507753eb..5b95b0f11d41cd 100644
--- a/clang/include/clang/Basic/DiagnosticSemaKinds.td
+++ b/clang/include/clang/Basic/DiagnosticSemaKinds.td
@@ -10171,9 +10171,6 @@ def err_fallthrough_attr_outside_switch : Error<
 def err_fallthrough_attr_invalid_placement : Error<
   "fallthrough annotation does not directly precede switch label">;
 
-def err_assume_attr_args : Error<
-  "attribute '%0' requires a single expression argument">;
-
 def warn_unreachable_default : Warning<
   "default label in switch which covers all enumeration values">,
   InGroup<CoveredSwitchDefault>, DefaultIgnore;
diff --git a/clang/lib/Parse/ParseDecl.cpp b/clang/lib/Parse/ParseDecl.cpp
index dd179414a14191..3423a1d7e22430 100644
--- a/clang/lib/Parse/ParseDecl.cpp
+++ b/clang/lib/Parse/ParseDecl.cpp
@@ -629,6 +629,9 @@ void Parser::ParseGNUAttributeArgs(
     ParseAttributeWithTypeArg(*AttrName, AttrNameLoc, Attrs, ScopeName,
                               ScopeLoc, Form);
     return;
+  } else if (AttrKind == ParsedAttr::AT_CXXAssume) {
+    ParseCXXAssumeAttributeArg(Attrs, AttrName, AttrNameLoc, EndLoc);
+    return;
   }
 
   // These may refer to the function arguments, but need to be parsed early to
@@ -683,6 +686,10 @@ unsigned Parser::ParseClangAttributeArgs(
     ParseTypeTagForDatatypeAttribute(*AttrName, AttrNameLoc, Attrs, EndLoc,
                                      ScopeName, ScopeLoc, Form);
     break;
+
+  case ParsedAttr::AT_CXXAssume:
+    ParseCXXAssumeAttributeArg(Attrs, AttrName, AttrNameLoc, EndLoc);
+    break;
   }
   return !Attrs.empty() ? Attrs.begin()->getNumArgs() : 0;
 }
diff --git a/clang/lib/Sema/SemaStmtAttr.cpp b/clang/lib/Sema/SemaStmtAttr.cpp
index 691857e88beb49..c72af6a24033d2 100644
--- a/clang/lib/Sema/SemaStmtAttr.cpp
+++ b/clang/lib/Sema/SemaStmtAttr.cpp
@@ -656,7 +656,8 @@ bool Sema::CheckRebuiltStmtAttributes(ArrayRef<const Attr *> Attrs) {
 ExprResult Sema::ActOnCXXAssumeAttr(Stmt *St, const ParsedAttr &A,
                                     SourceRange Range) {
   if (A.getNumArgs() != 1 || !A.getArgAsExpr(0)) {
-    Diag(A.getLoc(), diag::err_assume_attr_args) << A.getAttrName() << Range;
+    Diag(A.getLoc(), diag::err_attribute_wrong_number_arguments)
+        << A.getAttrName() << 1 << Range;
     return ExprError();
   }
 
diff --git a/clang/test/CodeGen/assume_attr.c b/clang/test/CodeGen/assume_attr.c
deleted file mode 100644
index 338a625188af08..00000000000000
--- a/clang/test/CodeGen/assume_attr.c
+++ /dev/null
@@ -1,58 +0,0 @@
-// RUN: %clang_cc1 -emit-llvm -triple i386-linux-gnu %s -o - | FileCheck %s
-// RUN: %clang_cc1 -x c -emit-pch -o %t %s
-// RUN: %clang_cc1 -include-pch %t %s -emit-llvm -o - | FileCheck %s
-
-// TODO: for "foo" and "bar", "after" is not added as it appears "after" the first use or definition respectively. There might be a way to allow that.
-
-// CHECK:   define{{.*}} void @bar() #0
-// CHECK:   define{{.*}} void @baz() #1
-// CHECK:   declare{{.*}} void @foo() #2
-// CHECK:      attributes #0
-// CHECK-SAME:   "llvm.assume"="bar:before1,bar:before2,bar:before3,bar:def1,bar:def2"
-// CHECK:      attributes #1
-// CHECK-SAME:   "llvm.assume"="baz:before1,baz:before2,baz:before3,baz:def1,baz:def2,baz:after"
-// CHECK:      attributes #2
-// CHECK-SAME:   "llvm.assume"="foo:before1,foo:before2,foo:before3"
-
-#ifndef HEADER
-#define HEADER
-
-/// foo: declarations only
-
-__attribute__((assume("foo:before1"))) void foo(void);
-
-__attribute__((assume("foo:before2")))
-__attribute__((assume("foo:before3"))) void
-foo(void);
-
-/// baz: static function declarations and a definition
-
-__attribute__((assume("baz:before1"))) static void baz(void);
-
-__attribute__((assume("baz:before2")))
-__attribute__((assume("baz:before3"))) static void
-baz(void);
-
-// Definition
-__attribute__((assume("baz:def1,baz:def2"))) static void baz(void) { foo(); }
-
-__attribute__((assume("baz:after"))) static void baz(void);
-
-/// bar: external function declarations and a definition
-
-__attribute__((assume("bar:before1"))) void bar(void);
-
-__attribute__((assume("bar:before2")))
-__attribute__((assume("bar:before3"))) void
-bar(void);
-
-// Definition
-__attribute__((assume("bar:def1,bar:def2"))) void bar(void) { baz(); }
-
-__attribute__((assume("bar:after"))) void bar(void);
-
-/// back to foo
-
-__attribute__((assume("foo:after"))) void foo(void);
-
-#endif
diff --git a/clang/test/CodeGenCXX/assume_attr.cpp b/clang/test/CodeGenCXX/assume_attr.cpp
index dbe76501377c0a..962dcc470f676a 100644
--- a/clang/test/CodeGenCXX/assume_attr.cpp
+++ b/clang/test/CodeGenCXX/assume_attr.cpp
@@ -8,77 +8,77 @@
 
 /// foo: declarations only
 
-__attribute__((assume("foo:before1"))) void foo();
+[[omp::assume("foo:before1")]] void foo();
 
-__attribute__((assume("foo:before2")))
-__attribute__((assume("foo:before3"))) void
+[[omp::assume("foo:before2")]]
+[[omp::assume("foo:before3")]] void
 foo();
 
 /// baz: static function declarations and a definition
 
-__attribute__((assume("baz:before1"))) static void baz();
+[[omp::assume("baz:before1")]] static void baz();
 
-__attribute__((assume("baz:before2")))
-__attribute__((assume("baz:before3"))) static void
+[[omp::assume("baz:before2")]]
+[[omp::assume("baz:before3")]] static void
 baz();
 
 // Definition
-__attribute__((assume("baz:def1,baz:def2"))) static void baz() { foo(); }
+[[omp::assume("baz:def1,baz:def2")]] static void baz() { foo(); }
 
-__attribute__((assume("baz:after"))) static void baz();
+[[omp::assume("baz:after")]] static void baz();
 
 /// bar: external function declarations and a definition
 
-__attribute__((assume("bar:before1"))) void bar();
+[[omp::assume("bar:before1")]] void bar();
 
-__attribute__((assume("bar:before2")))
-__attribute__((assume("bar:before3"))) void
+[[omp::assume("bar:before2")]]
+[[omp::assume("bar:before3")]] void
 bar();
 
 // Definition
-__attribute__((assume("bar:def1,bar:def2"))) void bar() { baz(); }
+[[omp::assume("bar:def1,bar:def2")]] void bar() { baz(); }
 
-__attribute__((assume("bar:after"))) void bar();
+[[omp::assume("bar:after")]] void bar();
 
 /// back to foo
 
-__attribute__((assume("foo:after"))) void foo();
+[[omp::assume("foo:after")]] void foo();
 
 /// class tests
 class C {
-  __attribute__((assume("C:private_method"))) void private_method();
-  __attribute__((assume("C:private_static"))) static void private_static();
+  [[omp::assume("C:private_method")]] void private_method();
+  [[omp::assume("C:private_static")]] static void private_static();
 
 public:
-  __attribute__((assume("C:public_method1"))) void public_method();
-  __attribute__((assume("C:public_static1"))) static void public_static();
+  [[omp::assume("C:public_method1")]] void public_method();
+  [[omp::assume("C:public_static1")]] static void public_static();
 };
 
-__attribute__((assume("C:public_method2"))) void C::public_method() {
+[[omp::assume("C:public_method2")]] void C::public_method() {
   private_method();
 }
 
-__attribute__((assume("C:public_static2"))) void C::public_static() {
+[[omp::assume("C:public_static2")]] void C::public_static() {
   private_static();
 }
 
 /// template tests
 template <typename T>
-__attribute__((assume("template_func<T>"))) void template_func() {}
+[[omp::assume("template_func<T>")]] void template_func() {}
 
 template <>
-__attribute__((assume("template_func<float>"))) void template_func<float>() {}
+[[omp::assume("template_func<float>")]] void template_func<float>() {}
 
 template <>
 void template_func<int>() {}
 
 template <typename T>
 struct S {
-  __attribute__((assume("S<T>::method"))) void method();
+  [[omp::assume("S<T>::method")]] void method();
 };
 
 template <>
-__attribute__((assume("S<float>::method"))) void S<float>::method() {}
+[[omp::assume("S<float>::method")]] void S<float>::method() {}
 
 template <>
 void S<int>::method() {}
diff --git a/clang/test/OpenMP/assumes_codegen.cpp b/clang/test/OpenMP/assumes_codegen.cpp
index 6a5871c303aade..2682e394d0d324 100644
--- a/clang/test/OpenMP/assumes_codegen.cpp
+++ b/clang/test/OpenMP/assumes_codegen.cpp
@@ -67,46 +67,46 @@ int lambda_outer() {
 }
 #pragma omp end assumes
 
-// AST:      __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) void foo() {
-// AST-NEXT: }
-// AST-NEXT: class BAR {
-// AST-NEXT: public:
-// AST-NEXT:     __attribute__((assume("ompx_range_bar_only"))) __attribute__((assume("ompx_range_bar_only_2"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) BAR()      {
-// AST-NEXT:     }
-// AST-NEXT:     __attribute__((assume("ompx_range_bar_only"))) __attribute__((assume("ompx_range_bar_only_2"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) void bar1()     {
-// AST-NEXT:     }
-// AST-NEXT:     __attribute__((assume("ompx_range_bar_only"))) __attribute__((assume("ompx_range_bar_only_2"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) static void bar2()      {
-// AST-NEXT:     }
-// AST-NEXT: };
-// AST-NEXT:  __attribute__((assume("ompx_range_bar_only"))) __attribute__((assume("ompx_range_bar_only_2"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) void bar() {
-// AST-NEXT:     BAR b;
-// AST-NEXT: }
-// AST-NEXT: void baz() __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp")));
-// AST-NEXT: template <typename T> class BAZ {
-// AST-NEXT: public:
-// AST-NEXT:     __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) BAZ<T>()      {
-// AST-NEXT:     }
-// AST-NEXT:     __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp")))  void baz1()     {
-// AST-NEXT:     }
-// AST-NEXT:     __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) static void baz2()      {
-// AST-NEXT:     }
-// AST-NEXT: };
-// AST-NEXT: template<> class BAZ<float> {
-// AST-NEXT: public:
-// AST-NEXT:     __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp")))  BAZ()    {
-// AST-NEXT:     }
-// AST-NEXT:     void baz1() __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp")));
-// AST-NEXT:     static void baz2() __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp")));
-// AST-NEXT: };
-// AST-NEXT: __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) void baz() {
-// AST-NEXT:     BAZ<float> b;
-// AST-NEXT: }
-// AST-NEXT: __attribute__((assume("ompx_lambda_assumption"))) __attribute__((assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses"))) __attribute__((assume("omp_no_openmp"))) int lambda_outer() {
-// AST-NEXT:     auto lambda_inner = []() {
-// AST-NEXT:         return 42;
-// AST-NEXT:     };
-// AST-NEXT:     return lambda_inner();
-// AST-NEXT: }
+// AST{LITERAL}:      void foo() [[omp::assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses")]] [[omp::assume("omp_no_openmp")]] {
+// AST-NEXT{LITERAL}: }
+// AST-NEXT{LITERAL}: class BAR {
+// AST-NEXT{LITERAL}: public:
+// AST-NEXT{LITERAL}:     BAR() [[omp::assume("ompx_range_bar_only")]] [[omp::assume("ompx_range_bar_only_2")]] [[omp::assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses")]] [[omp::assume("omp_no_openmp")]] {
+// AST-NEXT{LITERAL}:     }
+// AST-NEXT{LITERAL}:     void bar1() [[omp::assume("ompx_range_bar_only")]] [[omp::assume("ompx_range_bar_only_2")]] [[omp::assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses")]] [[omp::assume("omp_no_openmp")]] {
+// AST-NEXT{LITERAL}:     }
+// AST-NEXT{LITERAL}:     static void bar2() [[omp::assume("ompx_range_bar_only")]] [[omp::assume("ompx_range_bar_only_2")]] [[omp::assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses")]] [[omp::assume("omp_no_openmp")]] {
+// AST-NEXT{LITERAL}:     }
+// AST-NEXT{LITERAL}: };
+// AST-NEXT{LITERAL}: void bar() [[omp::assume("ompx_range_bar_only")]] [[omp::assume("ompx_range_bar_only_2")]] [[omp::assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses")]] [[omp::assume("omp_no_openmp")]] {
+// AST-NEXT{LITERAL}:     BAR b;
+// AST-NEXT{LITERAL}: }
+// AST-NEXT{LITERAL}: void baz() [[omp::assume("ompx_1234")]] [[omp::assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses")]] [[omp::assume("omp_no_openmp")]];
+// AST-NEXT{LITERAL}: template <typename T> class BAZ {
+// AST-NEXT{LITERAL}: public:
+// AST-NEXT{LITERAL}:     BAZ<T>() [[omp::assume("ompx_1234")]] [[omp::assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses")]] [[omp::assume("omp_no_openmp")]] {
+// AST-NEXT{LITERAL}:     }
+// AST-NEXT{LITERAL}:     void baz1() [[omp::assume("ompx_1234")]] [[omp::assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses")]] [[omp::assume("omp_no_openmp")]] {
+// AST-NEXT{LITERAL}:     }
+// AST-NEXT{LITERAL}:     static void baz2() [[omp::assume("ompx_1234")]] [[omp::assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses")]] [[omp::assume("omp_no_openmp")]] {
+// AST-NEXT{LITERAL}:     }
+// AST-NEXT{LITERAL}: };
+// AST-NEXT{LITERAL}: template<> class BAZ<float> {
+// AST-NEXT{LITERAL}: public:
+// AST-NEXT{LITERAL}:     BAZ() [[omp::assume("ompx_1234")]] [[omp::assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses")]] [[omp::assume("omp_no_openmp")]] [[omp::assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses")]] [[omp::assume("omp_no_openmp")]] {
+// AST-NEXT{LITERAL}:     }
+// AST-NEXT{LITERAL}:     void baz1() [[omp::assume("ompx_1234")]] [[omp::assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses")]] [[omp::assume("omp_no_openmp")]] [[omp::assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses")]] [[omp::assume("omp_no_openmp")]];
+// AST-NEXT{LITERAL}:     static void baz2() [[omp::assume("ompx_1234")]] [[omp::assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses")]] [[omp::assume("omp_no_openmp")]] [[omp::assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses")]] [[omp::assume("omp_no_openmp")]];
+// AST-NEXT{LITERAL}: };
+// AST-NEXT{LITERAL}: void baz() [[omp::assume("ompx_1234")]] [[omp::assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses")]] [[omp::assume("omp_no_openmp")]] {
+// AST-NEXT{LITERAL}:     BAZ<float> b;
+// AST-NEXT{LITERAL}: }
+// AST-NEXT{LITERAL}: int lambda_outer() [[omp::assume("ompx_lambda_assumption")]] [[omp::assume("omp_no_openmp_routines,ompx_another_warning,ompx_after_invalid_clauses")]] [[omp::assume("omp_no_openmp")]] {
+// AST-NEXT{LITERAL}:     auto lambda_inner = []() {
+// AST-NEXT{LITERAL}:         return 42;
+// AST-NEXT{LITERAL}:     };
+// AST-NEXT{LITERAL}:     return lambda_inner();
+// AST-NEXT{LITERAL}: }
 
 #endif
 
diff --git a/clang/test/OpenMP/assumes_print.cpp b/clang/test/OpenMP/assumes_print.cpp
index a7f04edb3b1af4..6f83912a277d29 100644
--- a/clang/test/OpenMP/assumes_print.cpp
+++ b/clang/test/OpenMP/assumes_print.cpp
@@ -37,8 +37,8 @@ void baz() {
 }
 #pragma omp end assumes
 
-// CHECK: __attribute__((assume("omp_no_openmp_routines"))) __attribute__((assume("omp_no_openmp"))) void foo()
-// CHECK: __attribute__((assume("ompx_range_bar_only"))) __attribute__((assume("ompx_range_bar_only_2"))) __attribute__((assume("omp_no_openmp_routines"))) __attribute__((assume("omp_no_openmp"))) void bar()
-// CHECK: __attribute__((assume("ompx_1234"))) __attribute__((assume("omp_no_openmp_routines"))) __attribute__((assume("omp_no_openmp"))) void baz()
+// CHECK{LITERAL}: void foo() [[omp::assume("omp_no_openmp_routines")]] [[omp::assume("omp_no_openmp")]]
+// CHECK{LITERAL}: void bar() [[omp::assume("ompx_range_bar_only")]] [[omp::assume("ompx_range_bar_only_2")]] [[omp::assume("omp_no_openmp_routines")]] [[omp::assume("omp_no_openmp")]]
+// CHECK{LITERAL}: void baz() [[omp::assume("ompx_1234")]] [[omp::assume("omp_no_openmp_routines")]] [[omp::assume("omp_no_openmp")]]
 
 #endif
diff --git a/clang/test/OpenMP/assumes_template_print.cpp b/clang/test/OpenMP/assumes_template_print.cpp
index bd1100fbefffc6..b1f3df7cb6fc6a 100644
--- a/clang/test/OpenMP/assumes_template_print.cpp
+++ b/clang/test/OpenMP/assumes_template_print.cpp
@@ -17,7 +17,7 @@ template <typename T>
 struct S {
   int a;
 // CHECK: template <typename T> struct S {
-// CHECK:     __attribute__((assume("ompx_global_assumption"))) void foo()     {
+// CHECK{LITERAL}:     void foo() [[omp::assume("ompx_global_assumption")]] {
   void foo() {
     #pragma omp parallel
     {}
@@ -25,15 +25,15 @@ struct S {
 };
 
 // CHECK: template<> struct S<int> {
-// CHECK:     __attribute__((assume("ompx_global_assumption"))) void foo()     {
+// CHECK{LITERAL}:     void foo() [[omp::assume("ompx_global_assumption")]] {
 
 #pragma omp begin assumes no_openmp
-// CHECK: __attribute__((assume("omp_no_openmp"))) __attribute__((assume("ompx_global_assumption"))) void S_with_assumes_no_call() {
+// CHECK{LITERAL}: void S_with_assumes_no_call() [[omp::assume("omp_no_openmp")]] [[omp::assume("ompx_global_assumption")]] {
 void S_with_assumes_no_call() {
   S<int> s;
   s.a = 0;
 }
-// CHECK: __attribute__((assume("omp_no_openmp"))) __attribute__((assume("ompx_global_assumption"))) void S_with_assumes_call() {
+// CHECK{LITERAL}: void S_with_assumes_call() [[omp::assume("omp_no_openmp")]] [[omp::assume("ompx_global_assumption")]] {
 void S_with_assumes_call() {
   S<int> s;
   s.a = 0;
@@ -42,7 +42,7 @@ void S_with_assumes_call() {
 }
 #pragma omp end assumes
 
-// CHECK: __attribute__((assume("ompx_global_assumption"))) void S_without_assumes() {
+// CHECK{LITERAL}: void S_without_assumes() [[omp::assume("ompx_global_assumption")]] {
 void S_without_assumes() {
   S<int> s;
   s.foo();
@@ -54,7 +54,7 @@ void S_without_assumes() {
 template <typename T>
 struct P {
 // CHECK: template <typename T> struct P {
-// CHECK:    __attribute__((assume("ompx_global_assumption"))) void foo()      {
+// CHECK{LITERAL}:    void foo() [[omp::assume("ompx_global_assumption")]] {
   int a;
   void foo() {
     #pragma omp parallel
@@ -65,21 +65,21 @@ struct P {
 // TODO: Avoid the duplication here:
 
 // CHECK: template<> struct P<int> {
-// CHECK:      __attribute__((assume("ompx_global_assumption"))) __attribute__((assume("ompx_global_assumption"))) void foo()   {
+// CHECK{LITERAL}:     void foo() [[omp::assume("ompx_global_assumption")]] [[omp::assume("ompx_global_assumption")]] {
 
-// CHECK: __attribute__((assume("ompx_global_assumption"))) void P_without_assumes() {
+// CHECK{LITERAL}: void P_without_assumes() [[omp::assume("ompx_global_assumption")]] {
 void P_without_assumes() {
   P<int> p;
   p.foo();
 }
 
 #pragma omp begin assumes no_openmp
-// CHECK: __attribute__((assume("omp_no_openmp"))) __attribute__((assume("ompx_global_assumption"))) void P_with_assumes_no_call() {
+// CHECK{LITERAL}: void P_with_assumes_no_call() [[omp::assume("omp_no_openmp")]] [[omp::assume("ompx_global_assumption")]] {
 void P_with_assumes_no_call() {
   P<int> p;
   p.a = 0;
 }
-// CHECK: __attribute__((assume("omp_no_openmp"))) __attribute__((assume("ompx_global_assumption"))) void P_with_assumes_call() {
+// CHECK{LITERAL}: void P_with_assumes_call() [[omp::assume("omp_no_openmp")]] [[omp::assume("ompx_global_assumption")]] {
 void P_with_assumes_call() {
   P<int> p;
   p.a = 0;
diff --git a/clang/test/OpenMP/remarks_parallel_in_multiple_target_state_machines.c b/clang/test/OpenMP/remarks_parallel_in_multiple_target_state_machines.c
index 2f829d2ad0945e..1afedc6683f86d 100644
--- a/clang/test/OpenMP/remarks_parallel_in_multiple_target_state_machines.c
+++ b/clang/test/OpenMP/remarks_parallel_in_multiple_target_state_machines.c
@@ -4,7 +4,7 @@
 
 // host-no-diagnostics
 
-void baz(void) __attribute__((assume("omp_no_openmp")));
+[[omp::assume("omp_no_openmp")]] void baz(void);
 
 void bar1(void) {
 #pragma omp parallel // #0
@@ -24,7 +24,7 @@ void foo1(void) {
                          // all-remark@#2 {{Rewriting generic-mode kernel with a customized state machine. [OMP131]}}
 
   {
-    baz();           // all-remark {{Value has potential side effects preventing SPMD-mode execution. Add `__attribute__((assume("ompx_spmd_amenable")))` to the called function to override. [OMP121]}}
+    baz();           // all-remark {{Value has potential side effects preventing SPMD-mode execution. Add `[[omp::assume("ompx_spmd_amenable")]]` to the called function to override. [OMP121]}}
 #pragma omp parallel // #3
     {
     }
@@ -39,7 +39,7 @@ void foo2(void) {
 #pragma omp target teams // #5
                          // all-remark@#5 {{Rewriting generic-mode kernel with a customized state machine. [OMP131]}}
   {
-    baz();           // all-remark {{Value has potential side effects preventing SPMD-mode execution. Add `__attribute__((assume("ompx_spmd_amenable")))` to the called function to override. [OMP121]}}
+    baz();           // all-remark {{Value has potential side effects preventing SPMD-mode execution. Add `[[omp::assume("ompx_spmd_amenable")]]` to the called function to override. [OMP121]}}
 #pragma omp parallel // #6
     {
     }
@@ -57,7 +57,7 @@ void foo3(void) {
 #pragma omp target teams // #8
                          // all-remark@#8 {{Rewriting generic-mode kernel with a customized state machine. [OMP131]}}
   {
-    baz();           // all-remark {{Value has potential side effects preventing SPMD-mode execution. Add `__attribute__((assume("ompx_spmd_amenable")))` to the called function to override. [OMP121]}}
+    baz();           // all-remark {{Value has potential side effects preventing SPMD-mode execution. Add `[[omp::assume("ompx_spmd_amenable")]]` to the called function to override. [OMP121]}}
 #pragma omp parallel // #9
     {
     }
diff --git a/clang/test/OpenMP/remarks_parallel_in_target_state_machine.c b/clang/test/OpenMP/remarks_parallel_in_target_state_machine.c
index c48a4b966077d6..5ce8f1fa4046d9 100644
--- a/clang/test/OpenMP/remarks_parallel_in_target_state_machine.c
+++ b/clang/test/OpenMP/remarks_parallel_in_target_state_machine.c
@@ -3,7 +3,7 @@
 
 // host-no-diagnostics
 
-void baz(void) __attribute__((assume("omp_no_openmp")));
+[[omp::assume("omp_no_openmp")]] void baz(void);
 
 void bar(void) {
 #pragma omp parallel // #1                                                                                                                                                                                                                                                                                                                                           \
@@ -16,7 +16,7 @@ void foo(void) {
 #pragma omp target teams // #2
                          // expected-remark@#2 {{Rewriting generic-mode kernel with a customized state machine. [OMP131]}}
   {
-    baz();               // expected-remark {{Value has potential side effects preventing SPMD-mode execution. Add `__attribute__((assume("ompx_spmd_amenable")))` to the called function to override. [OMP121]}}
+    baz();               // expected-remark {{Value has potential side effects preventing SPMD-mode execution. Add `[[omp::assume("ompx_spmd_amenable")]]` to the called function to override. [OMP121]}}
 #pragma omp parallel
     {
     }
diff --git a/clang/test/Sema/attr-assume.c b/clang/test/Sema/attr-assume.c
deleted file mode 100644
index 98deffa3a74609..00000000000000
--- a/clang/test/Sema/attr-assume.c
+++ /dev/null
@@ -1,14 +0,0 @@
-// RUN: %clang_cc1 -triple i386-apple-darwin9 -fsyntax-only -verify %s
-
-void f1(void) __attribute__((assume(3))); // expected-error {{expected string literal as argument of 'assume' attribute}}
-void f2(void) __attribute__((assume(int))); // expected-error {{expected string literal as argument of 'assume' attribute}}
-void f3(void) __attribute__((assume(for))); // expected-error {{expected string literal as argument of 'assume' attribute}}
-void f4(void) __attribute__((assume("QQQQ"))); // expected-warning {{unknown assumption string 'QQQQ'; attribute is potentially ignored}}
-void f5(void) __attribute__((assume("omp_no_openmp")));
-void f6(void) __attribute__((assume("omp_noopenmp"))); // expected-warning {{unknown assumption string 'omp_noopenmp' may be misspelled; attribute is potentially ignored, did you mean 'omp_no_openmp'?}}
-void f7(void) __attribute__((assume("omp_no_openmp_routine"))); // expected-warning {{unknown assumption string 'omp_no_openmp_routine' may be misspelled; attribute is potentially ignored, did you mean 'omp_no_openmp_routines'?}}
-void f8(void) __attribute__((assume("omp_no_openmp1"))); // expected-warning {{unknown assumption string 'omp_no_openmp1' may be misspelled; attribute is potentially ignored, did you mean 'omp_no_openmp'?}}
-void f9(void) __attribute__((assume("omp_no_openmp", "omp_no_openmp"))); // expected-error {{'assume' attribute takes one argument}}
-
-int g1 __attribute__((assume(0))); // expected-error {{expected string literal as argument of 'assume' attribute}}
-int g2 __attribute__((assume("omp_no_openmp"))); // expected-warning {{'assume' attribute only applies to functions and Objective-C methods}}
diff --git a/clang/test/SemaCXX/cxx23-assume.cpp b/clang/test/SemaCXX/cxx23-assume.cpp
index 2d7c9b174d9019..2f9a6390ed0262 100644
--- a/clang/test/SemaCXX/cxx23-assume.cpp
+++ b/clang/test/SemaCXX/cxx23-assume.cpp
@@ -56,6 +56,10 @@ void g(int x) {
   [[assume(true)]] while (false) {} // expected-error {{only applies to empty statements}}
   [[assume(true)]] label:; // expected-error {{cannot be applied to a declaration}}
   [[assume(true)]] goto label; // expected-error {{only applies to empty statements}}
+
+  // Also check variant spellings.
+  __attribute__((assume(true))) {}; // expected-error {{only applies to empty statements}}
+  [[clang::assume(true)]] {}; // expected-error {{only applies to empty statements}}
 }
 
 // Check that 'x' is ODR-used here.
diff --git a/libclc/generic/include/clc/clcfunc.h b/libclc/generic/include/clc/clcfunc.h
index ad9eb23f293331..fa6c094db2a14d 100644
--- a/libclc/generic/include/clc/clcfunc.h
+++ b/libclc/generic/include/clc/clcfunc.h
@@ -8,7 +8,7 @@
 #define _CLC_DEF
 #elif defined(CLC_CLSPV) || defined(CLC_CLSPV64)
 #define _CLC_DEF                                                               \
-  __attribute__((noinline)) __attribute__((assume("clspv_libclc_builtin")))
+  __attribute__((noinline)) __attribute__((__omp_assume__("clspv_libclc_builtin")))
 #else
 #define _CLC_DEF __attribute__((always_inline))
 #endif
diff --git a/llvm/lib/Transforms/IPO/OpenMPOpt.cpp b/llvm/lib/Transforms/IPO/OpenMPOpt.cpp
index eea9399127e8ee..e3a4821b8226b4 100644
--- a/llvm/lib/Transforms/IPO/OpenMPOpt.cpp
+++ b/llvm/lib/Transforms/IPO/OpenMPOpt.cpp
@@ -4238,7 +4238,7 @@ struct AAKernelInfoFunction : AAKernelInfo {
           ORA << "Value has potential side effects preventing SPMD-mode "
                  "execution";
           if (isa<CallBase>(NonCompatibleI)) {
-            ORA << ". Add `__attribute__((assume(\"ompx_spmd_amenable\")))` to "
+            ORA << ". Add `[[omp::assume(\"ompx_spmd_amenable\")]]` to "
                    "the called function to override";
           }
           return ORA << ".";
@@ -4380,7 +4380,7 @@ struct AAKernelInfoFunction : AAKernelInfo {
           continue;
         auto Remark = [&](OptimizationRemarkAnalysis ORA) {
           return ORA << "Call may contain unknown parallel regions. Use "
-                     << "`__attribute__((assume(\"omp_no_parallelism\")))` to "
+                     << "`[[omp::assume(\"omp_no_parallelism\")]]` to "
                         "override.";
         };
         A.emitRemark<OptimizationRemarkAnalysis>(UnknownParallelRegionCB,
diff --git a/llvm/test/Transforms/OpenMP/custom_state_machines.ll b/llvm/test/Transforms/OpenMP/custom_state_machines.ll
index b0d1842ca2e8ba..310851b57cbf3f 100644
--- a/llvm/test/Transforms/OpenMP/custom_state_machines.ll
+++ b/llvm/test/Transforms/OpenMP/custom_state_machines.ll
@@ -8,7 +8,7 @@
 ;; void p1(void);
 ;; int unknown(void);
 ;; void unknown_pure(void) __attribute__((pure));
-;; void unknown_no_openmp(void) __attribute__((assume("omp_no_openmp")));
+;; [[omp::assume("omp_no_openmp")]] void unknown_no_openmp(void);
 ;;
 ;; int G;
 ;; void no_parallel_region_in_here(void) {
diff --git a/llvm/test/Transforms/OpenMP/custom_state_machines_pre_lto.ll b/llvm/test/Transforms/OpenMP/custom_state_machines_pre_lto.ll
index fe134ce350dc83..cbb291d4e4079e 100644
--- a/llvm/test/Transforms/OpenMP/custom_state_machines_pre_lto.ll
+++ b/llvm/test/Transforms/OpenMP/custom_state_machines_pre_lto.ll
@@ -10,7 +10,7 @@
 ;; void p1(void);
 ;; int unknown(void);
 ;; void unknown_pure(void) __attribute__((pure));
-;; void unknown_no_openmp(void) __attribute__((assume("omp_no_openmp")));
+;; [[omp::assume("omp_no_openmp")]] void unknown_no_openmp(void);
 ;;
 ;; int G;
 ;; void no_parallel_region_in_here(void) {
diff --git a/llvm/test/Transforms/OpenMP/custom_state_machines_remarks.ll b/llvm/test/Transforms/OpenMP/custom_state_machines_remarks.ll
index f8c4e6b113c9b4..f7bfd306506944 100644
--- a/llvm/test/Transforms/OpenMP/custom_state_machines_remarks.ll
+++ b/llvm/test/Transforms/OpenMP/custom_state_machines_remarks.ll
@@ -1,10 +1,10 @@
 ; RUN: opt -passes=openmp-opt -pass-remarks=openmp-opt -pass-remarks-missed=openmp-opt -pass-remarks-analysis=openmp-opt -disable-output < %s 2>&1 | FileCheck %s
 target triple = "nvptx64"
 
-; CHECK: remark: llvm/test/Transforms/OpenMP/custom_state_machines_remarks.c:11:1: Generic-mode kernel is executed with a customized state machine that requires a fallback.
-; CHECK: remark: llvm/test/Transforms/OpenMP/custom_state_machines_remarks.c:13:5: Call may contain unknown parallel regions. Use `__attribute__((assume("omp_no_parallelism")))` to override.
-; CHECK: remark: llvm/test/Transforms/OpenMP/custom_state_machines_remarks.c:15:5: Call may contain unknown parallel regions. Use `__attribute__((assume("omp_no_parallelism")))` to override.
-; CHECK: remark: llvm/test/Transforms/OpenMP/custom_state_machines_remarks.c:20:1: Rewriting generic-mode kernel with a customized state machine.
+; CHECK{LITERAL}: remark: llvm/test/Transforms/OpenMP/custom_state_machines_remarks.c:11:1: Generic-mode kernel is executed with a customized state machine that requires a fallback.
+; CHECK{LITERAL}: remark: llvm/test/Transforms/OpenMP/custom_state_machines_remarks.c:13:5: Call may contain unknown parallel regions. Use `[[omp::assume("omp_no_parallelism")]]` to override.
+; CHECK{LITERAL}: remark: llvm/test/Transforms/OpenMP/custom_state_machines_remarks.c:15:5: Call may contain unknown parallel regions. Use `[[omp::assume("omp_no_parallelism")]]` to override.
+; CHECK{LITERAL}: remark: llvm/test/Transforms/OpenMP/custom_state_machines_remarks.c:20:1: Rewriting generic-mode kernel with a customized state machine.
 
 
 ;; void unknown(void);
@@ -24,7 +24,7 @@ target triple = "nvptx64"
 ;;   }
 ;; }
 ;;
-;; void no_openmp(void) __attribute__((assume("omp_no_openmp")));
+;; [[omp::assume("omp_no_openmp")]] void no_openmp(void);
 ;; void test_no_fallback(void) {
 ;;   #pragma omp target teams
 ;;   {
diff --git a/llvm/test/Transforms/OpenMP/spmdization.ll b/llvm/test/Transforms/OpenMP/spmdization.ll
index ef95aed24ded44..034caa2929112e 100644
--- a/llvm/test/Transforms/OpenMP/spmdization.ll
+++ b/llvm/test/Transforms/OpenMP/spmdization.ll
@@ -7,7 +7,7 @@
 ; RUN: opt --mtriple=nvptx64-- -S -passes=openmp-opt-postlink < %s | FileCheck %s --check-prefix=NVPTX-DISABLED2
 
 ;; void unknown(void);
-;; void spmd_amenable(void) __attribute__((assume("ompx_spmd_amenable")));
+;; [[omp::assume("ompx_spmd_amenable")]] void spmd_amenable(void);
 ;;
 ;; void sequential_loop() {
 ;;   #pragma omp target teams
@@ -22,7 +22,7 @@
 ;;   }
 ;; }
 ;;
-;; void use(__attribute__((noescape)) int *) __attribute__((assume("ompx_spmd_amenable")));
+;; [[omp::assume("ompx_spmd_amenable")]] void use(__attribute__((noescape)) int *);
 ;;
 ;; void sequential_loop_to_stack_var() {
 ;;   #pragma omp target teams
diff --git a/llvm/test/Transforms/OpenMP/spmdization_guarding.ll b/llvm/test/Transforms/OpenMP/spmdization_guarding.ll
index 45e86e4ce2efb0..9b9716f513ec3e 100644
--- a/llvm/test/Transforms/OpenMP/spmdization_guarding.ll
+++ b/llvm/test/Transforms/OpenMP/spmdization_guarding.ll
@@ -2,8 +2,8 @@
 ; RUN: opt -S -passes=openmp-opt < %s | FileCheck %s
 ; RUN: opt -S -passes=openmp-opt -openmp-opt-disable-spmdization < %s | FileCheck %s --check-prefix=CHECK-DISABLED
 ;
-;    void pure(void) __attribute__((pure, assume("ompx_spmd_amenable")));
-;    int no_openmp(int *) __attribute__((assume("omp_no_openmp","ompx_spmd_amenable")));
+;    [[omp::assume("ompx_spmd_amenable")]] void pure(void) __attribute__((pure));
+;    [[omp::assume("omp_no_openmp","ompx_spmd_amenable")]] int no_openmp(int *);
 ;
 ;    void sequential_loop(int *x, int N) {
 ;    #pragma omp target teams
diff --git a/llvm/test/Transforms/OpenMP/spmdization_remarks.ll b/llvm/test/Transforms/OpenMP/spmdization_remarks.ll
index 28df2f52491308..f5a4cea9a841cb 100644
--- a/llvm/test/Transforms/OpenMP/spmdization_remarks.ll
+++ b/llvm/test/Transforms/OpenMP/spmdization_remarks.ll
@@ -1,12 +1,12 @@
 ; RUN: opt -passes=openmp-opt -pass-remarks=openmp-opt -pass-remarks-missed=openmp-opt -pass-remarks-analysis=openmp-opt -disable-output < %s 2>&1 | FileCheck %s
 target triple = "nvptx64"
 
-; CHECK: remark: llvm/test/Transforms/OpenMP/spmdization_remarks.c:13:5: Value has potential side effects preventing SPMD-mode execution. Add `__attribute__((assume("ompx_spmd_amenable")))` to the called function to override.
-; CHECK: remark: llvm/test/Transforms/OpenMP/spmdization_remarks.c:15:5: Value has potential side effects preventing SPMD-mode execution. Add `__attribute__((assume("ompx_spmd_amenable")))` to the called function to override.
-; CHECK: remark: llvm/test/Transforms/OpenMP/spmdization_remarks.c:11:1: Generic-mode kernel is executed with a customized state machine that requires a fallback.
-; CHECK: remark: llvm/test/Transforms/OpenMP/spmdization_remarks.c:13:5: Call may contain unknown parallel regions. Use `__attribute__((assume("omp_no_parallelism")))` to override.
-; CHECK: remark: llvm/test/Transforms/OpenMP/spmdization_remarks.c:15:5: Call may contain unknown parallel regions. Use `__attribute__((assume("omp_no_parallelism")))` to override.
-; CHECK: remark: llvm/test/Transforms/OpenMP/spmdization_remarks.c:20:1: Transformed generic-mode kernel to SPMD-mode.
+; CHECK{LITERAL}: remark: llvm/test/Transforms/OpenMP/spmdization_remarks.c:13:5: Value has potential side effects preventing SPMD-mode execution. Add `[[omp::assume("ompx_spmd_amenable")]]` to the called function to override.
+; CHECK{LITERAL}: remark: llvm/test/Transforms/OpenMP/spmdization_remarks.c:15:5: Value has potential side effects preventing SPMD-mode execution. Add `[[omp::assume("ompx_spmd_amenable")]]` to the called function to override.
+; CHECK{LITERAL}: remark: llvm/test/Transforms/OpenMP/spmdization_remarks.c:11:1: Generic-mode kernel is executed with a customized state machine that requires a fallback.
+; CHECK{LITERAL}: remark: llvm/test/Transforms/OpenMP/spmdization_remarks.c:13:5: Call may contain unknown parallel regions. Use `[[omp::assume("omp_no_parallelism")]]` to override.
+; CHECK{LITERAL}: remark: llvm/test/Transforms/OpenMP/spmdization_remarks.c:15:5: Call may contain unknown parallel regions. Use `[[omp::assume("omp_no_parallelism")]]` to override.
+; CHECK{LITERAL}: remark: llvm/test/Transforms/OpenMP/spmdization_remarks.c:20:1: Transformed generic-mode kernel to SPMD-mode.
 
 
 ;; void unknown(void);
@@ -26,7 +26,7 @@ target triple = "nvptx64"
 ;;   }
 ;; }
 ;;
-;; void no_openmp(void) __attribute__((assume("omp_no_openmp")));
+;; void no_openmp(void) [[omp::assume("omp_no_openmp")]];
 ;; void test_no_fallback(void) {
 ;;   #pragma omp target teams
 ;;   {
diff --git a/openmp/docs/remarks/OMP121.rst b/openmp/docs/remarks/OMP121.rst
index 88561b8a1fe1af..f3ceeac7f3ab92 100644
--- a/openmp/docs/remarks/OMP121.rst
+++ b/openmp/docs/remarks/OMP121.rst
@@ -1,6 +1,6 @@
 .. _omp121:
 
-Value has potential side effects preventing SPMD-mode execution. Add `__attribute__((assume(\"ompx_spmd_amenable\")))` to the called function to override. [OMP121]
+Value has potential side effects preventing SPMD-mode execution. Add `[[omp::assume(\"ompx_spmd_amenable\")]]` to the called function to override. [OMP121]
 ===================================================================================================================================================================
 
 This analysis remarks indicates that a potential side-effect that cannot be
@@ -42,7 +42,7 @@ or operations that cannot be executed in SPMD-mode.
 
    $ clang++ -fopenmp -fopenmp-targets=nvptx64 -O2 -Rpass-analysis=openmp-opt omp121.cpp
    omp121.cpp:8:13: remark: Value has potential side effects preventing SPMD-mode
-   execution.  Add `__attribute__((assume("ompx_spmd_amenable")))` to the called function
+   execution.  Add `[[omp::assume("ompx_spmd_amenable")]]` to the called function
    to override. [OMP121]
    int x = work();
             ^
@@ -53,7 +53,7 @@ contain any code that prevents SPMD-mode execution.
 
 .. code-block:: c++
 
-  __attribute__((assume("ompx_spmd_amenable"))) extern int work();
+  [[omp::assume("ompx_spmd_amenable")]] extern int work();
 
   void use(int x);
 
diff --git a/openmp/docs/remarks/OMP133.rst b/openmp/docs/remarks/OMP133.rst
index f025352de105b8..5a734479d495f2 100644
--- a/openmp/docs/remarks/OMP133.rst
+++ b/openmp/docs/remarks/OMP133.rst
@@ -1,4 +1,4 @@
-Call may contain unknown parallel regions. Use `__attribute__((assume("omp_no_parallelism")))` to override. [OMP133]
+Call may contain unknown parallel regions. Use `[[omp::assume("omp_no_parallelism")]]` to override. [OMP133]
 ====================================================================================================================
 
 .. _omp133:
@@ -33,7 +33,7 @@ regions. This is typically coupled with the :ref:`OMP132 <omp132>` remark.
 
    $ clang++ -fopenmp -fopenmp-targets=nvptx64 -O2 -Rpass-analysis=openmp-opt omp133.cpp
    omp133.cpp:6:5: remark: Call may contain unknown parallel regions. Use
-   `__attribute__((assume("omp_no_parallelism")))` to override. [OMP133]
+   `[[omp::assume("omp_no_parallelism")]]` to override. [OMP133]
    setup();
    ^
 
@@ -43,7 +43,7 @@ specialized state machine.
 
 .. code-block:: c++
 
-   __attribute__((assume("omp_no_parallelism"))) extern void setup();
+   [[omp::assume("omp_no_parallelism")]] extern void setup();
 
 
    void foo() {
diff --git a/openmp/docs/remarks/OptimizationRemarks.rst b/openmp/docs/remarks/OptimizationRemarks.rst
index a29dce60e073c6..2c683a4376c491 100644
--- a/openmp/docs/remarks/OptimizationRemarks.rst
+++ b/openmp/docs/remarks/OptimizationRemarks.rst
@@ -81,7 +81,7 @@ OpenMP Remarks
    * - :ref:`OMP121 <omp121>`
      - Analysis
      - Value has potential side effects preventing SPMD-mode execution. Add
-       `__attribute__((assume(\"ompx_spmd_amenable\")))` to the called function
+       `[[omp::assume(\"ompx_spmd_amenable\")]]` to the called function
        to override.
    * - :ref:`OMP130 <omp130>`
      - Optimization
@@ -96,7 +96,7 @@ OpenMP Remarks
    * - :ref:`OMP133 <omp133>`
      - Analysis
      - Call may contain unknown parallel regions. Use
-       `__attribute__((assume("omp_no_parallelism")))` to override.
+       `[[omp::assume("omp_no_parallelism")]]` to override.
    * - :ref:`OMP140 <omp140>`
      - Analysis
      - Could not internalize function. Some optimizations may not be possible.



More information about the cfe-commits mailing list