[clang] 0a41fb7 - [Clang] Treat `ext_vector_type` as a regular type attribute (#130177)

[via cfe-commits]

Author: Joseph Huber
Date: 2025-03-07T10:09:06-06:00
New Revision: 0a41fb71f12624f470b8489f997cbe43b188e0a1

URL: https://github.com/llvm/llvm-project/commit/0a41fb71f12624f470b8489f997cbe43b188e0a1
DIFF: https://github.com/llvm/llvm-project/commit/0a41fb71f12624f470b8489f997cbe43b188e0a1.diff

LOG: [Clang] Treat `ext_vector_type` as a regular type attribute (#130177)

Summary:
This attribute is mostly borrowed from OpenCL, but is useful in general
for accessing the LLVM vector types. Previously the only way to use it
was through typedefs. This patch changes that to allow use as a regular
type attribute, similar to address spaces.

Added: 
    clang/test/Sema/vector-ast.cpp

Modified: 
    clang/docs/ReleaseNotes.rst
    clang/include/clang/Basic/Attr.td
    clang/include/clang/Basic/AttrDocs.td
    clang/lib/Sema/SemaDeclAttr.cpp
    clang/test/CodeGen/vector.c
    clang/test/CodeGenCUDA/amdgpu-bf16.cu
    clang/test/Sema/types.c

Removed: 
    


################################################################################
diff  --git a/clang/docs/ReleaseNotes.rst b/clang/docs/ReleaseNotes.rst
index 577b3f2130df7..7d443b4ee8a29 100644
--- a/clang/docs/ReleaseNotes.rst
+++ b/clang/docs/ReleaseNotes.rst
@@ -151,6 +151,7 @@ Adding [[clang::unsafe_buffer_usage]] attribute to a method definition now turns
 related warnings within the method body.
 
 - The ``no_sanitize`` attribute now accepts both ``gnu`` and ``clang`` names.
+- The ``ext_vector_type(n)`` attribute can now be used as a generic type attribute.
 - Clang now diagnoses use of declaration attributes on void parameters. (#GH108819)
 - Clang now allows ``__attribute__((model("small")))`` and
   ``__attribute__((model("large")))`` on non-TLS globals in x86-64 compilations.

diff  --git a/clang/include/clang/Basic/Attr.td b/clang/include/clang/Basic/Attr.td
index ccfe69d32e0a6..8d7a9ef0cb1cd 100644
--- a/clang/include/clang/Basic/Attr.td
+++ b/clang/include/clang/Basic/Attr.td
@@ -1721,17 +1721,10 @@ def EnableIf : InheritableAttr {
   let Documentation = [EnableIfDocs];
 }
 
-def ExtVectorType : Attr {
-  // This is an OpenCL-related attribute and does not receive a [[]] spelling.
-  let Spellings = [GNU<"ext_vector_type">];
-  // FIXME: This subject list is wrong; this is a type attribute.
-  let Subjects = SubjectList<[TypedefName], ErrorDiag>;
+def ExtVectorType : TypeAttr {
+  let Spellings = [Clang<"ext_vector_type">];
   let Args = [ExprArgument<"NumElements">];
-  let ASTNode = 0;
-  let Documentation = [Undocumented];
-  // This is a type attribute with an incorrect subject list, so should not be
-  // permitted by #pragma clang attribute.
-  let PragmaAttributeSupport = 0;
+  let Documentation = [ExtVectorTypeDocs];
 }
 
 def FallThrough : StmtAttr {

diff  --git a/clang/include/clang/Basic/AttrDocs.td b/clang/include/clang/Basic/AttrDocs.td
index fdc58d1c92c0d..a1b08ac23ee9d 100644
--- a/clang/include/clang/Basic/AttrDocs.td
+++ b/clang/include/clang/Basic/AttrDocs.td
@@ -1113,6 +1113,42 @@ template instantiation, so the value for ``T::number`` is known.
   }];
 }
 
+def ExtVectorTypeDocs : Documentation {
+  let Category = DocCatFunction;
+  let Content = [{
+The ext_vector_type(N) attribute specifies that a type is a vector with N
+elements, directly mapping to an LLVM vector type. Originally from OpenCL, it
+allows element access the array subscript operator ``[]``, ``sN`` where ``N`` is
+a hexadecimal value, or ``x``, ``y``, ``z``, ``w`` for graphics-style indexing.
+This attribute enables efficient SIMD operations and is usable in
+general-purpose code.
+
+.. code-block:: c++
+
+  template <typename T, uint32_t N>
+  constexpr T simd_reduce(T [[clang::ext_vector_type(N)]] v) {
+    static_assert((N & (N - 1)) == 0, "N must be a power of two");
+    if constexpr (N == 1) {
+      return v[0];
+    } else {
+      T [[clang::ext_vector_type(N / 2)]] reduced = v.hi + v.lo;
+      return simd_reduce(reduced);
+    }
+  }
+
+The vector type also supports swizzling up to sixteen elements. This can be done
+using the object accessors. The OpenCL documentation lists the full list of
+accepted values.
+.. code-block:: c++
+
+  using f16_x16 = _Float16 __attribute__((ext_vector_type(16)));
+
+  f16_x16 reverse(f16_x16 v) { return v.sfedcba9876543210; }
+
+See the OpenCL documentation for some more complete examples.
+  }];
+}
+
 def DiagnoseIfDocs : Documentation {
   let Category = DocCatFunction;
   let Content = [{

diff  --git a/clang/lib/Sema/SemaDeclAttr.cpp b/clang/lib/Sema/SemaDeclAttr.cpp
index c1e97aa2dde5b..de2696ea62e3b 100644
--- a/clang/lib/Sema/SemaDeclAttr.cpp
+++ b/clang/lib/Sema/SemaDeclAttr.cpp
@@ -1191,7 +1191,8 @@ static void handleTestTypestateAttr(Sema &S, Decl *D, const ParsedAttr &AL) {
 
 static void handleExtVectorTypeAttr(Sema &S, Decl *D, const ParsedAttr &AL) {
   // Remember this typedef decl, we will need it later for diagnostics.
-  S.ExtVectorDecls.push_back(cast<TypedefNameDecl>(D));
+  if (isa<TypedefNameDecl>(D))
+    S.ExtVectorDecls.push_back(cast<TypedefNameDecl>(D));
 }
 
 static void handlePackedAttr(Sema &S, Decl *D, const ParsedAttr &AL) {

diff  --git a/clang/test/CodeGen/vector.c b/clang/test/CodeGen/vector.c
index c20e320463056..5d677aaf16948 100644
--- a/clang/test/CodeGen/vector.c
+++ b/clang/test/CodeGen/vector.c
@@ -8,6 +8,10 @@ void test1(void) {
 __v4hi x = {1,2,3};
 __v4hi y = {1,2,3,4};
 
+
+// CHECK: @z = local_unnamed_addr global <8 x float> zeroinitializer
+float z __attribute__((ext_vector_type(8)));
+
 typedef int vty __attribute((vector_size(16)));
 int test2(void) { vty b; return b[2LL]; }
 
@@ -18,9 +22,6 @@ void test3 ( vec4* a, char b, float c ) {
   (*a)[b] = c;
 }
 
-
-
-
 #include <mmintrin.h>
 
 int test4(int argc, char *argv[]) {

diff  --git a/clang/test/CodeGenCUDA/amdgpu-bf16.cu b/clang/test/CodeGenCUDA/amdgpu-bf16.cu
index 4610b4ae3cbe5..f9b067d3fe0d3 100644
--- a/clang/test/CodeGenCUDA/amdgpu-bf16.cu
+++ b/clang/test/CodeGenCUDA/amdgpu-bf16.cu
@@ -111,19 +111,15 @@ __device__ __bf16 test_call( __bf16 in) {
 // CHECK-NEXT:    ret void
 //
 __device__ void test_vec_assign() {
-  typedef __attribute__((ext_vector_type(2))) __bf16 bf16_x2;
-  bf16_x2 vec2_a, vec2_b;
+  __bf16 [[clang::ext_vector_type(2)]] vec2_a, vec2_b;
   vec2_a = vec2_b;
 
-  typedef __attribute__((ext_vector_type(4))) __bf16 bf16_x4;
-  bf16_x4 vec4_a, vec4_b;
+  __bf16 __attribute__((ext_vector_type(4))) vec4_a, vec4_b;
   vec4_a = vec4_b;
 
-  typedef __attribute__((ext_vector_type(8))) __bf16 bf16_x8;
-  bf16_x8 vec8_a, vec8_b;
+  __bf16 [[clang::ext_vector_type(8)]] vec8_a, vec8_b;
   vec8_a = vec8_b;
 
-  typedef __attribute__((ext_vector_type(16))) __bf16 bf16_x16;
-  bf16_x16 vec16_a, vec16_b;
+  __bf16 __attribute__((ext_vector_type(16))) vec16_a, vec16_b;
   vec16_a = vec16_b;
 }

diff  --git a/clang/test/Sema/types.c b/clang/test/Sema/types.c
index 2a5f530740e9a..2be0e6544f3d7 100644
--- a/clang/test/Sema/types.c
+++ b/clang/test/Sema/types.c
@@ -78,7 +78,7 @@ typedef int __attribute__((ext_vector_type(0))) e4;                // expected-e
 // no support for vector enum type
 enum { e_2 } x3 __attribute__((vector_size(64))); // expected-error {{invalid vector element type}}
 
-int x4 __attribute__((ext_vector_type(64)));  // expected-error {{'ext_vector_type' attribute only applies to typedefs}}
+int x4 __attribute__((ext_vector_type(64)));
 
 typedef __attribute__ ((ext_vector_type(32),__aligned__(32))) unsigned char uchar32;
 

diff  --git a/clang/test/Sema/vector-ast.cpp b/clang/test/Sema/vector-ast.cpp
new file mode 100644
index 0000000000000..c9823719feabe
--- /dev/null
+++ b/clang/test/Sema/vector-ast.cpp
@@ -0,0 +1,15 @@
+// RUN: %clang_cc1 %s -verify -ast-dump | FileCheck %s
+
+// expected-no-diagnostics
+
+//      CHECK: ExtVectorType {{.*}} 'int __attribute__((ext_vector_type(4)))' 4
+// CHECK-NEXT: BuiltinType {{.*}} 'int'
+int x __attribute__((ext_vector_type(4)));
+using ExtVecType = decltype(x);
+
+// CHECK: FunctionDecl {{.*}} 'int () __attribute__((ext_vector_type(4)))'
+int __attribute__((ext_vector_type(4))) foo() { return x; }
+//      CHECK:  CompoundStmt
+// CHECK-NEXT: ReturnStmt
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'int __attribute__((ext_vector_type(4)))' <LValueToRValue>
+// CHECK-NEXT: DeclRefExpr {{.*}} 'int __attribute__((ext_vector_type(4)))' lvalue Var {{.*}} 'x' 'int __attribute__((ext_vector_type(4)))'