[HLSL][RootSignature] Define and integrate rootsig clang attr and decl

clang/include/clang/AST/Decl.h

@@ -41,6 +41,7 @@
 #include "llvm/ADT/PointerUnion.h"
 #include "llvm/ADT/StringRef.h"
 #include "llvm/ADT/iterator_range.h"
+#include "llvm/Frontend/HLSL/HLSLRootSignature.h"
 #include "llvm/Support/Casting.h"
 #include "llvm/Support/Compiler.h"
 #include "llvm/Support/TrailingObjects.h"
@@ -5178,6 +5179,42 @@ class HLSLBufferDecl final : public NamedDecl, public DeclContext {
   friend class ASTDeclWriter;
 };
 
+class HLSLRootSignatureDecl final
+    : public NamedDecl,
+      private llvm::TrailingObjects<HLSLRootSignatureDecl,
+                                    llvm::hlsl::rootsig::RootElement> {
+  friend TrailingObjects;
+
+  unsigned NumElems;
+
+  llvm::hlsl::rootsig::RootElement *getElems() {
+    return getTrailingObjects<llvm::hlsl::rootsig::RootElement>();
+  }
+
+  const llvm::hlsl::rootsig::RootElement *getElems() const {
+    return getTrailingObjects<llvm::hlsl::rootsig::RootElement>();
+  }
+
+  HLSLRootSignatureDecl(DeclContext *DC, SourceLocation Loc, IdentifierInfo *ID,
+                        unsigned NumElems);
+
+public:
+  static HLSLRootSignatureDecl *
+  Create(ASTContext &C, DeclContext *DC, SourceLocation Loc, IdentifierInfo *ID,
+         ArrayRef<llvm::hlsl::rootsig::RootElement> RootElements);
+
+  static HLSLRootSignatureDecl *CreateDeserialized(ASTContext &C,
+                                                   GlobalDeclID ID);
+
+  ArrayRef<llvm::hlsl::rootsig::RootElement> getRootElements() const {
+    return {getElems(), NumElems};
+  }
+
+  // Implement isa/cast/dyncast/etc.
+  static bool classof(const Decl *D) { return classofKind(D->getKind()); }
+  static bool classofKind(Kind K) { return K == HLSLRootSignature; }
+};
+
 /// Insertion operator for diagnostics.  This allows sending NamedDecl's
 /// into a diagnostic with <<.
 inline const StreamingDiagnostic &operator<<(const StreamingDiagnostic &PD,

clang/include/clang/AST/RecursiveASTVisitor.h

@@ -1599,6 +1599,8 @@ DEF_TRAVERSE_DECL(EmptyDecl, {})
 
 DEF_TRAVERSE_DECL(HLSLBufferDecl, {})
 
+DEF_TRAVERSE_DECL(HLSLRootSignatureDecl, {})
+
 DEF_TRAVERSE_DECL(LifetimeExtendedTemporaryDecl, {
   TRY_TO(TraverseStmt(D->getTemporaryExpr()));
 })

clang/include/clang/AST/TextNodeDumper.h

@@ -408,6 +408,7 @@ class TextNodeDumper
   void
   VisitLifetimeExtendedTemporaryDecl(const LifetimeExtendedTemporaryDecl *D);
   void VisitHLSLBufferDecl(const HLSLBufferDecl *D);
+  void VisitHLSLRootSignatureDecl(const HLSLRootSignatureDecl *D);
   void VisitHLSLOutArgExpr(const HLSLOutArgExpr *E);
   void VisitOpenACCConstructStmt(const OpenACCConstructStmt *S);
   void VisitOpenACCLoopConstruct(const OpenACCLoopConstruct *S);

clang/include/clang/Basic/Attr.td

@@ -4735,6 +4735,17 @@ def Error : InheritableAttr {
   let Documentation = [ErrorAttrDocs];
 }
 
+/// HLSL Root Signature Attribute
+def RootSignature : Attr {
+  /// [RootSignature(Signature)]
+  let Spellings = [Microsoft<"RootSignature">];
+  let Args = [IdentifierArgument<"Signature">];
+  let Subjects = SubjectList<[Function],
+                             ErrorDiag, "'function'">;
+  let LangOpts = [HLSL];
+  let Documentation = [RootSignatureDocs];
+}
+
 def HLSLNumThreads: InheritableAttr {
   let Spellings = [Microsoft<"numthreads">];
   let Args = [IntArgument<"X">, IntArgument<"Y">, IntArgument<"Z">];

clang/include/clang/Basic/AttrDocs.td

@@ -8195,6 +8195,17 @@ and https://microsoft.github.io/hlsl-specs/proposals/0013-wave-size-range.html
   }];
 }
 
+def RootSignatureDocs : Documentation {
+  let Category = DocCatFunction;
+  let Content = [{
+The ``RootSignature`` attribute applies to HLSL entry functions to define what
+types of resources are bound to the graphics pipeline.
+
+For details about the use and specification of Root Signatures please see here:
+https://learn.microsoft.com/en-us/windows/win32/direct3d12/root-signatures
+  }];
+}
+
 def NumThreadsDocs : Documentation {
   let Category = DocCatFunction;
   let Content = [{

clang/include/clang/Basic/DeclNodes.td

@@ -111,5 +111,6 @@ def Empty : DeclNode<Decl>;
 def RequiresExprBody : DeclNode<Decl>, DeclContext;
 def LifetimeExtendedTemporary : DeclNode<Decl>;
 def HLSLBuffer : DeclNode<Named, "HLSLBuffer">, DeclContext;
+def HLSLRootSignature : DeclNode<Named, "HLSLRootSignature">;
 def OpenACCDeclare : DeclNode<Decl, "#pragma acc declare">;
 def OpenACCRoutine : DeclNode<Decl, "#pragma acc routine">;

clang/include/clang/Parse/Parser.h

@@ -3093,6 +3093,7 @@ class Parser : public CodeCompletionHandler {
     return AttrsParsed;
   }
   void ParseMicrosoftUuidAttributeArgs(ParsedAttributes &Attrs);
+  void ParseMicrosoftRootSignatureAttributeArgs(ParsedAttributes &Attrs);
   void ParseMicrosoftAttributes(ParsedAttributes &Attrs);
   bool MaybeParseMicrosoftDeclSpecs(ParsedAttributes &Attrs) {
     if (getLangOpts().DeclSpecKeyword && Tok.is(tok::kw___declspec)) {

clang/include/clang/Sema/SemaHLSL.h

@@ -119,6 +119,7 @@ class SemaHLSL : public SemaBase {
                                        bool IsCompAssign);
   void emitLogicalOperatorFixIt(Expr *LHS, Expr *RHS, BinaryOperatorKind Opc);
 
+  void handleRootSignatureAttr(Decl *D, const ParsedAttr &AL);
   void handleNumThreadsAttr(Decl *D, const ParsedAttr &AL);
   void handleWaveSizeAttr(Decl *D, const ParsedAttr &AL);
   void handleSV_DispatchThreadIDAttr(Decl *D, const ParsedAttr &AL);

clang/lib/AST/CMakeLists.txt

@@ -2,6 +2,7 @@ set(LLVM_LINK_COMPONENTS
   BinaryFormat
   Core
   FrontendOpenMP
+  FrontendHLSL
   Support
   TargetParser
   )

clang/lib/AST/Decl.cpp

@@ -5847,6 +5847,38 @@ bool HLSLBufferDecl::buffer_decls_empty() {
   return DefaultBufferDecls.empty() && decls_empty();
 }
 
+//===----------------------------------------------------------------------===//
+// HLSLRootSignatureDecl Implementation
+//===----------------------------------------------------------------------===//
+
+HLSLRootSignatureDecl::HLSLRootSignatureDecl(DeclContext *DC,
+                                             SourceLocation Loc,
+                                             IdentifierInfo *ID,
+                                             unsigned NumElems)
+    : NamedDecl(Decl::Kind::HLSLRootSignature, DC, Loc, DeclarationName(ID)),
+      NumElems(NumElems) {}
+
+HLSLRootSignatureDecl *HLSLRootSignatureDecl::Create(
+    ASTContext &C, DeclContext *DC, SourceLocation Loc, IdentifierInfo *ID,
+    ArrayRef<llvm::hlsl::rootsig::RootElement> RootElements) {
+  HLSLRootSignatureDecl *RSDecl =
+      new (C, DC,
+           additionalSizeToAlloc<llvm::hlsl::rootsig::RootElement>(
+               RootElements.size()))
+          HLSLRootSignatureDecl(DC, Loc, ID, RootElements.size());
+  auto *StoredElems = RSDecl->getElems();
+  std::uninitialized_copy(RootElements.begin(), RootElements.end(),
+                          StoredElems);
+  return RSDecl;
+}
+
+HLSLRootSignatureDecl *
+HLSLRootSignatureDecl::CreateDeserialized(ASTContext &C, GlobalDeclID ID) {
+  HLSLRootSignatureDecl *Result = new (C, ID)
+      HLSLRootSignatureDecl(nullptr, SourceLocation(), nullptr, /*NumElems=*/0);
+  return Result;
+}
+
 //===----------------------------------------------------------------------===//
 // ImportDecl Implementation
 //===----------------------------------------------------------------------===//

clang/lib/AST/DeclBase.cpp

@@ -886,6 +886,7 @@ unsigned Decl::getIdentifierNamespaceForKind(Kind DeclKind) {
     case ObjCProperty:
     case MSProperty:
     case HLSLBuffer:
+    case HLSLRootSignature:
       return IDNS_Ordinary;
     case Label:
       return IDNS_Label;

clang/lib/AST/TextNodeDumper.cpp

@@ -24,6 +24,7 @@
 #include "clang/Basic/Specifiers.h"
 #include "clang/Basic/TypeTraits.h"
 #include "llvm/ADT/StringExtras.h"
+#include "llvm/Frontend/HLSL/HLSLRootSignature.h"
 
 #include <algorithm>
 #include <utility>
@@ -3037,6 +3038,12 @@ void TextNodeDumper::VisitHLSLBufferDecl(const HLSLBufferDecl *D) {
   dumpName(D);
 }
 
+void TextNodeDumper::VisitHLSLRootSignatureDecl(
+    const HLSLRootSignatureDecl *D) {
+  dumpName(D);
+  llvm::hlsl::rootsig::dumpRootElements(OS, D->getRootElements());
+}
+
 void TextNodeDumper::VisitHLSLOutArgExpr(const HLSLOutArgExpr *E) {
   OS << (E->isInOut() ? " inout" : " out");
 }

clang/lib/CodeGen/CGDecl.cpp

@@ -106,6 +106,7 @@ void CodeGenFunction::EmitDecl(const Decl &D, bool EvaluateConditionDecl) {
   case Decl::Binding:
   case Decl::UnresolvedUsingIfExists:
   case Decl::HLSLBuffer:
+  case Decl::HLSLRootSignature:
     llvm_unreachable("Declaration should not be in declstmts!");
   case Decl::Record:    // struct/union/class X;
   case Decl::CXXRecord: // struct/union/class X; [C++]

clang/lib/Parse/ParseDeclCXX.cpp

@@ -21,10 +21,12 @@
 #include "clang/Basic/TargetInfo.h"
 #include "clang/Basic/TokenKinds.h"
 #include "clang/Lex/LiteralSupport.h"
+#include "clang/Parse/ParseHLSLRootSignature.h"
 #include "clang/Parse/Parser.h"
 #include "clang/Parse/RAIIObjectsForParser.h"
 #include "clang/Sema/DeclSpec.h"
 #include "clang/Sema/EnterExpressionEvaluationContext.h"
+#include "clang/Sema/Lookup.h"
 #include "clang/Sema/ParsedTemplate.h"
 #include "clang/Sema/Scope.h"
 #include "clang/Sema/SemaCodeCompletion.h"
@@ -5311,6 +5313,90 @@ void Parser::ParseMicrosoftUuidAttributeArgs(ParsedAttributes &Attrs) {
   }
 }
 
+void Parser::ParseMicrosoftRootSignatureAttributeArgs(ParsedAttributes &Attrs) {
+  assert(Tok.is(tok::identifier) &&
+         "Expected an identifier to denote which MS attribute to consider");
+  IdentifierInfo *RootSignatureIdent = Tok.getIdentifierInfo();
+  assert(RootSignatureIdent->getName() == "RootSignature" &&
+         "Expected RootSignature identifier for root signature attribute");
+
+  SourceLocation RootSignatureLoc = Tok.getLocation();
+  ConsumeToken();
+
+  // Ignore the left paren location for now.
+  BalancedDelimiterTracker T(*this, tok::l_paren);
+  if (T.consumeOpen()) {
+    Diag(Tok, diag::err_expected) << tok::l_paren;
+    return;
+  }
+
+  auto ProcessStringLiteral = [this]() -> std::optional<StringLiteral *> {
+    if (!isTokenStringLiteral())
+      return std::nullopt;
+
+    ExprResult StringResult = ParseUnevaluatedStringLiteralExpression();
+    if (StringResult.isInvalid())
+      return std::nullopt;
+
+    if (auto Lit = dyn_cast<StringLiteral>(StringResult.get()))
+      return Lit;
+
+    return std::nullopt;
+  };
+
+  auto StrLiteral = ProcessStringLiteral();
+  if (!StrLiteral.has_value()) {
+    Diag(Tok, diag::err_expected_string_literal)
+        << /*in attributes...*/ 4 << RootSignatureIdent->getName();
+    SkipUntil(tok::r_paren, StopAtSemi | StopBeforeMatch);
+    T.consumeClose();
+    return;
+  }
+
+  // Construct our identifier
+  StringRef Signature = StrLiteral.value()->getString();
+  auto Hash = llvm::hash_value(Signature);
+  std::string IdStr = "__hlsl_rootsig_decl_" + std::to_string(Hash);
+  IdentifierInfo *DeclIdent = &(Actions.getASTContext().Idents.get(IdStr));
+
+  LookupResult R(Actions, DeclIdent, SourceLocation(),
+                 Sema::LookupOrdinaryName);
+  // Check if we have already found a decl of the same name, if we haven't
+  // then parse the root signature string and construct the in-memory elements
+  if (!Actions.LookupQualifiedName(R, Actions.CurContext)) {
+    SourceLocation SignatureLoc =
+        StrLiteral.value()->getExprLoc().getLocWithOffset(
+            1); // offset 1 for '"'
+    // Invoke the root signature parser to construct the in-memory constructs
+    hlsl::RootSignatureLexer Lexer(Signature, SignatureLoc);
+    SmallVector<llvm::hlsl::rootsig::RootElement> RootElements;
+    hlsl::RootSignatureParser Parser(RootElements, Lexer, PP);
+    if (Parser.parse()) {
+      T.consumeClose();
+      return;
+    }
+
+    // Create the Root Signature
+    auto *SignatureDecl = HLSLRootSignatureDecl::Create(
+        Actions.getASTContext(), /*DeclContext=*/Actions.CurContext,
+        RootSignatureLoc, DeclIdent, RootElements);
+    SignatureDecl->setImplicit();
+    Actions.PushOnScopeChains(SignatureDecl, getCurScope());
+  }
+
+  // Create the arg for the ParsedAttr
+  IdentifierLoc *ILoc = ::new (Actions.getASTContext())
+      IdentifierLoc(RootSignatureLoc, DeclIdent);
+
+  ArgsVector Args = {ILoc};
+
+  if (!T.consumeClose())
+    Attrs.addNew(RootSignatureIdent,
+                 SourceRange(RootSignatureLoc, T.getCloseLocation()), nullptr,
+                 SourceLocation(), Args.data(), Args.size(),
+                 ParsedAttr::Form::Microsoft());
+}
+
 /// ParseMicrosoftAttributes - Parse Microsoft attributes [Attr]
 ///
 /// [MS] ms-attribute:
@@ -5345,6 +5431,8 @@ void Parser::ParseMicrosoftAttributes(ParsedAttributes &Attrs) {
         break;
       if (Tok.getIdentifierInfo()->getName() == "uuid")
         ParseMicrosoftUuidAttributeArgs(Attrs);
+      else if (Tok.getIdentifierInfo()->getName() == "RootSignature")
+        ParseMicrosoftRootSignatureAttributeArgs(Attrs);
       else {
         IdentifierInfo *II = Tok.getIdentifierInfo();
         SourceLocation NameLoc = Tok.getLocation();

clang/lib/Sema/SemaDeclAttr.cpp

@@ -7481,6 +7481,9 @@ ProcessDeclAttribute(Sema &S, Scope *scope, Decl *D, const ParsedAttr &AL,
     break;
 
   // HLSL attributes:
+  case ParsedAttr::AT_RootSignature:
+    S.HLSL().handleRootSignatureAttr(D, AL);
+    break;
   case ParsedAttr::AT_HLSLNumThreads:
     S.HLSL().handleNumThreadsAttr(D, AL);
     break;

clang/lib/Sema/SemaHLSL.cpp

@@ -29,6 +29,7 @@
 #include "clang/Basic/Specifiers.h"
 #include "clang/Basic/TargetInfo.h"
 #include "clang/Sema/Initialization.h"
+#include "clang/Sema/Lookup.h"
 #include "clang/Sema/ParsedAttr.h"
 #include "clang/Sema/Sema.h"
 #include "clang/Sema/Template.h"
@@ -950,6 +951,33 @@ void SemaHLSL::emitLogicalOperatorFixIt(Expr *LHS, Expr *RHS,
       << NewFnName << FixItHint::CreateReplacement(FullRange, OS.str());
 }
 
+void SemaHLSL::handleRootSignatureAttr(Decl *D, const ParsedAttr &AL) {
+  if (AL.getNumArgs() != 1) {
+    Diag(AL.getLoc(), diag::err_attribute_wrong_number_arguments) << AL << 1;
+    return;
+  }
+
+  IdentifierInfo *Ident = AL.getArgAsIdent(0)->getIdentifierInfo();
+  if (auto *RS = D->getAttr<RootSignatureAttr>()) {
+    if (RS->getSignature() != Ident) {
+      Diag(AL.getLoc(), diag::err_disallowed_duplicate_attribute) << RS;
+      return;
+    }
+
+    Diag(AL.getLoc(), diag::warn_duplicate_attribute_exact) << RS;
+    return;
+  }
+
+  LookupResult R(SemaRef, Ident, SourceLocation(), Sema::LookupOrdinaryName);
+  if (SemaRef.LookupQualifiedName(R, D->getDeclContext()))
+    if (auto *SignatureDecl =
+            dyn_cast<HLSLRootSignatureDecl>(R.getFoundDecl())) {
+      // Perform validation of constructs here
+      D->addAttr(::new (getASTContext())
+                     RootSignatureAttr(getASTContext(), AL, Ident));
+    }
+}
+
 void SemaHLSL::handleNumThreadsAttr(Decl *D, const ParsedAttr &AL) {
   llvm::VersionTuple SMVersion =
       getASTContext().getTargetInfo().getTriple().getOSVersion();

clang/lib/Sema/SemaTemplateInstantiateDecl.cpp

@@ -999,6 +999,11 @@ Decl *TemplateDeclInstantiator::VisitHLSLBufferDecl(HLSLBufferDecl *Decl) {
   llvm_unreachable("HLSL buffer declarations cannot be instantiated");
 }
 
+Decl *TemplateDeclInstantiator::VisitHLSLRootSignatureDecl(
+    HLSLRootSignatureDecl *Decl) {
+  llvm_unreachable("HLSL root signature declarations cannot be instantiated");
+}
+
 Decl *
 TemplateDeclInstantiator::VisitPragmaCommentDecl(PragmaCommentDecl *D) {
   llvm_unreachable("pragma comment cannot be instantiated");

clang/lib/Serialization/ASTCommon.cpp

@@ -458,6 +458,7 @@ bool serialization::isRedeclarableDeclKind(unsigned Kind) {
   case Decl::RequiresExprBody:
   case Decl::UnresolvedUsingIfExists:
   case Decl::HLSLBuffer:
+  case Decl::HLSLRootSignature:
   case Decl::OpenACCDeclare:
   case Decl::OpenACCRoutine:
     return false;