https://github.com/vangthao95 created https://github.com/llvm/llvm-project/pull/214247
With -fms-extensions, a CUDA/HIP lambda capture list can be consumed as a Microsoft attribute list. This causes attributed lambdas used in direct initialization to be misparsed as function declarators. Tentatively inspect the tokens following the capture list and any trailing attributes to identify CUDA/HIP lambdas before parsing Microsoft attributes. Assisted by: Cursor / Claude Opus 4.8 >From 45060280c3e0e65d6d3a894355b8524bce545591 Mon Sep 17 00:00:00 2001 From: Vang Thao <[email protected]> Date: Tue, 4 Aug 2026 23:24:28 -0400 Subject: [PATCH] [clang][HIP][CUDA] Disambiguate lambdas from Microsoft attributes With -fms-extensions, a CUDA/HIP lambda capture list can be consumed as a Microsoft attribute list. This causes attributed lambdas used in direct initialization to be misparsed as function declarators. Tentatively inspect the tokens following the capture list and any trailing attributes to identify CUDA/HIP lambdas before parsing Microsoft attributes. Add host and device regression coverage for capture forms and preserve valid Microsoft attribute parsing. Assisted by: Cursor / Claude Opus 4.8 --- clang/include/clang/Parse/Parser.h | 10 +++- clang/lib/Parse/ParseExprCXX.cpp | 17 +++++++ clang/test/Parser/ms-lambda-capture.hip | 61 +++++++++++++++++++++++++ 3 files changed, 87 insertions(+), 1 deletion(-) create mode 100644 clang/test/Parser/ms-lambda-capture.hip diff --git a/clang/include/clang/Parse/Parser.h b/clang/include/clang/Parse/Parser.h index 08b67cb08cdd5..9b4cca05832de 100644 --- a/clang/include/clang/Parse/Parser.h +++ b/clang/include/clang/Parse/Parser.h @@ -2341,8 +2341,10 @@ class Parser : public CodeCompletionHandler { bool MaybeParseMicrosoftAttributes(ParsedAttributes &Attrs) { bool AttrsParsed = false; + // A '[' may begin a Microsoft attribute or a C++ lambda, so only parse it + // as an attribute after confirming it does not start a lambda. if ((getLangOpts().MicrosoftExt || getLangOpts().HLSL) && - Tok.is(tok::l_square)) { + Tok.is(tok::l_square) && !startsLambdaNotMicrosoftAttribute()) { ParsedAttributes AttrsWithRange(AttrFactory); ParseMicrosoftAttributes(AttrsWithRange); AttrsParsed = !AttrsWithRange.empty(); @@ -4718,6 +4720,12 @@ class Parser : public CodeCompletionHandler { /// If we are not looking at a lambda expression, returns ExprError(). ExprResult TryParseLambdaExpression(); + /// Returns true if the current '[' begins a CUDA/HIP lambda rather than a + /// Microsoft '[]' attribute (enabled under -fms-extensions or HLSL). + /// Restricted to CUDA/HIP, the only mode that allows attributes immediately + /// after a lambda's capture list. + bool startsLambdaNotMicrosoftAttribute(); + /// Parse a lambda introducer. /// \param Intro A LambdaIntroducer filled in with information about the /// contents of the lambda-introducer. diff --git a/clang/lib/Parse/ParseExprCXX.cpp b/clang/lib/Parse/ParseExprCXX.cpp index 06ff45da413ca..a50d968c2a550 100644 --- a/clang/lib/Parse/ParseExprCXX.cpp +++ b/clang/lib/Parse/ParseExprCXX.cpp @@ -748,6 +748,23 @@ ExprResult Parser::TryParseLambdaExpression() { return ParseLambdaExpressionAfterIntroducer(Intro); } +bool Parser::startsLambdaNotMicrosoftAttribute() { + // Restricted to CUDA/HIP, the only mode that allows attributes immediately + // after a lambda's capture list. + if (!getLangOpts().CUDA || Tok.isNot(tok::l_square)) + return false; + + // Skip the '[...]' and any trailing attributes (e.g. CUDA/HIP's + // '__device__'). A lambda then continues with '(', '{' or '<', while an + // attribute is followed by the declaration it applies to (e.g. '[propget] + // int get()'). + RevertingTentativeParsingAction TPA(*this); + ConsumeBracket(); + if (!SkipUntil(tok::r_square, StopAtSemi) || !TrySkipAttributes()) + return false; + return Tok.isOneOf(tok::l_paren, tok::l_brace, tok::less); +} + bool Parser::ParseLambdaIntroducer(LambdaIntroducer &Intro, LambdaIntroducerTentativeParse *Tentative) { if (Tentative) diff --git a/clang/test/Parser/ms-lambda-capture.hip b/clang/test/Parser/ms-lambda-capture.hip new file mode 100644 index 0000000000000..414913308c27a --- /dev/null +++ b/clang/test/Parser/ms-lambda-capture.hip @@ -0,0 +1,61 @@ +// RUN: %clang_cc1 -fms-extensions -DMS_EXTENSIONS -fsyntax-only -verify %s +// RUN: %clang_cc1 -fsyntax-only -verify %s +// RUN: %clang_cc1 -fcuda-is-device -fms-extensions -DMS_EXTENSIONS -fsyntax-only -verify %s +// RUN: %clang_cc1 -fcuda-is-device -fsyntax-only -verify %s +// expected-no-diagnostics + +// A '[' can begin a Microsoft attribute list under -fms-extensions, but it can +// also begin a C++ lambda-introducer. A lambda passed to a constructor via CTAD +// must not have its capture list consumed as a (possibly empty) Microsoft +// attribute, which previously misparsed the initializer as a function +// declarator under -fms-extensions. +// +// A lambda and a Microsoft attribute are told apart by what follows the +// (possibly attributed) '[...]' introducer. A lambda continues with a parameter +// list or body, whereas a Microsoft attribute is followed by the declaration +// it appertains to. + +#include "Inputs/cuda.h" + +template <typename F> struct Wrapper { + F f; + __host__ Wrapper(F fn) : f(fn) {} +}; +template <typename F> Wrapper(F) -> Wrapper<F>; + +void empty_and_defaults() { + Wrapper a([] __device__ () {}); + Wrapper b([=] __device__ () {}); + Wrapper c([&] __device__ () {}); + Wrapper d([] __device__ {}); + (void)a; (void)b; (void)c; (void)d; +} + +void named_captures() { + int values[] = {1}; + int x = 0; + Wrapper a([x] __device__ () { return x; }); + Wrapper b([&x] __device__ () { x++; }); + Wrapper c([y = x] __device__ () { return y; }); + Wrapper d([x, &x2 = x] __device__ () { return x + x2; }); + Wrapper e([a = values[0], b = [] { return 2; }()] __device__ () { + return a + b; + }); + (void)a; (void)b; (void)c; (void)d; (void)e; +} + +struct S { + int m = 0; + void g() { + Wrapper a([this] __device__ () { return m; }); + Wrapper b([*this] __device__ () { return m; }); + (void)a; (void)b; + } +}; + +#ifdef MS_EXTENSIONS +struct MicrosoftAttributeControls { + [] static int empty; + [propget] int get(); +}; +#endif _______________________________________________ cfe-commits mailing list [email protected] https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
