[clang] [CLANG] Enable alignas after GNU attributes (PR #133107)

via cfe-commits cfe-commits at lists.llvm.org
Wed Mar 26 08:30:58 PDT 2025


https://github.com/DenisGZM created https://github.com/llvm/llvm-project/pull/133107

Enable parsing alignas attribute after GNU attributes, before ParseDeclaration

This might be useful for cuda code where __shared__ and other specificators may be mixed with align.

I'd be glad to see if there are any better places or other technique to process this attribute without interrupting current flow of parsing.

>From c46eda67cd7434dcce5c1f29125a940dc4ff64ba Mon Sep 17 00:00:00 2001
From: Denis Gerasimov <dengzmm at gmail.com>
Date: Wed, 26 Mar 2025 18:29:38 +0300
Subject: [PATCH] [CLANG] Enable alignas after GNU attributes

---
 clang/lib/Parse/ParseStmt.cpp          |  5 +++++
 clang/test/SemaCUDA/cuda-attr-order.cu | 15 +++++++++++++++
 2 files changed, 20 insertions(+)
 create mode 100644 clang/test/SemaCUDA/cuda-attr-order.cu

diff --git a/clang/lib/Parse/ParseStmt.cpp b/clang/lib/Parse/ParseStmt.cpp
index 150b2879fc94f..33b9f63bcfa08 100644
--- a/clang/lib/Parse/ParseStmt.cpp
+++ b/clang/lib/Parse/ParseStmt.cpp
@@ -296,6 +296,11 @@ StmtResult Parser::ParseStatementOrDeclarationAfterAttributes(
     goto Retry;
   }
 
+  case tok::kw_alignas: {
+    ParseAlignmentSpecifier(CXX11Attrs);
+    goto Retry;
+  }
+
   case tok::kw_template: {
     SourceLocation DeclEnd;
     ParseTemplateDeclarationOrSpecialization(DeclaratorContext::Block, DeclEnd,
diff --git a/clang/test/SemaCUDA/cuda-attr-order.cu b/clang/test/SemaCUDA/cuda-attr-order.cu
new file mode 100644
index 0000000000000..d3bf5b014d1c6
--- /dev/null
+++ b/clang/test/SemaCUDA/cuda-attr-order.cu
@@ -0,0 +1,15 @@
+// Verify that we can parse a simple CUDA file with different attributes order.
+// RUN: %clang_cc1 "-triple" "nvptx-nvidia-cuda"  -fsyntax-only -verify %s
+// expected-no-diagnostics
+#include "Inputs/cuda.h"
+
+struct alignas(16) float4 {
+    float x, y, z, w;
+};
+
+__attribute__((device)) float func() {
+    __shared__ alignas(alignof(float4)) float As[4][4];  // Both combinations
+    alignas(alignof(float4)) __shared__  float Bs[4][4]; // must be legal
+
+    return As[0][0] + Bs[0][0];
+}



More information about the cfe-commits mailing list