[RFC][OpenCL][PoC] Testing TableGen with diffing

[azabaznov]

Summary

This adds an option string to clang-check to ignore certain AST node types
in order to achive suitable printing policy for comparison with output
generated from TableGen. Header-like TableGen generation is based on https://reviews.llvm.org/D97869.

[Anastasia]

This approach is really great because the testing is so quick! While it is not providing the coverage for invocation through clang, it can still be very valuable in various forms to enhance the testing and improve the coverage at very little cost. Thanks for working on this!

[azabaznov]

Thanks for feedback!

Locally I get really nice result (I am using Ubuntu 20.04):

$ llvm-lit  SemaOpenCL/compare-header-and-tablegen.cl

-- Testing: 1 tests, 1 workers --
FAIL: Clang :: SemaOpenCL/compare-header-and-tablegen.cl (1 of 1)
********************
Failed Tests (1):
  Clang :: SemaOpenCL/compare-header-and-tablegen.cl

Testing Time: 0.20s

I see no reason why it may take longer on other platforms.

clang/include/clang/AST/PrettyPrinter.h

PrintingPolicy(const LangOptions &LO)
         SuppressDefaultTemplateArgs(true), Bool(LO.Bool),
         Nullptr(LO.CPlusPlus11), Restrict(LO.C99), Alignof(LO.CPlusPlus11),
         UnderscoreAlignof(LO.C11), UseVoidForZeroParams(!LO.CPlusPlus),
         SplitTemplateClosers(!LO.CPlusPlus11), TerseOutput(false),
         PolishForDeclaration(false), Half(LO.Half),
         MSWChar(LO.MicrosoftExt && !LO.WChar), IncludeNewlines(true),
         MSVCFormatting(false), ConstantsAsWritten(false),
         SuppressImplicitBase(false), FullyQualifiedName(false),
-        PrintCanonicalTypes(false), PrintInjectedClassNameWithArguments(true) {}
+        PrintCanonicalTypes(false), PrintInjectedClassNameWithArguments(true),
+        SuppressRecordDecls(false), SuppressTypedefDecls(false),
+        SuppressEnumDecls(false), SuppressDeclNames(false) {}
 
   /// Adjust this printing policy for cases where it's known that we're
   /// printing C++ code (for instance, if AST dumping reaches a C++-only
   /// construct). This should not be used if a real LangOptions object is
   /// available.
   void adjustForCPlusPlus() {
     SuppressTagKeyword = true;
     Bool = true;
struct PrintingPolicy {
 
   /// Whether to print an InjectedClassNameType with template arguments or as
   /// written. When a template argument is unnamed, printing it results in
   /// invalid C++ code.
   unsigned PrintInjectedClassNameWithArguments : 1;
 
   /// Callbacks to use to allow the behavior of printing to be customized.
   const PrintingCallbacks *Callbacks = nullptr;
+
+  unsigned SuppressRecordDecls : 1;
+  unsigned SuppressTypedefDecls : 1;
+  unsigned SuppressEnumDecls : 1;
+  unsigned SuppressDeclNames : 1;
 };
 
 } // end namespace clang
 
 #endif

clang/include/clang/Frontend/ASTConsumers.h

 class LangOptions;
 class Preprocessor;
 class TargetOptions;
 
 // AST pretty-printer: prints out the AST in a format that is close to the
 // original C code.  The output is intended to be in a format such that
 // clang could re-parse the output back into the same AST, but the
 // implementation is still incomplete.
-std::unique_ptr<ASTConsumer> CreateASTPrinter(std::unique_ptr<raw_ostream> OS,
-                                              StringRef FilterString);
+std::unique_ptr<ASTConsumer>
+CreateASTPrinter(std::unique_ptr<raw_ostream> OS, StringRef FilterString,
+                 StringRef FilterNodeTypeString = "");
 
 // AST dumper: dumps the raw AST in human-readable form to the given output
 // stream, or stdout if OS is nullptr.
 std::unique_ptr<ASTConsumer>
 CreateASTDumper(std::unique_ptr<raw_ostream> OS, StringRef FilterString,
                 bool DumpDecls, bool Deserialize, bool DumpLookups,
-                bool DumpDeclTypes, ASTDumpOutputFormat Format);
+                bool DumpDeclTypes, ASTDumpOutputFormat Format,
+                StringRef FilterNodeTypeString = "");
 
 // AST Decl node lister: prints qualified names of all filterable AST Decl
 // nodes.
 std::unique_ptr<ASTConsumer> CreateASTDeclNodeLister();
 
 // Graphical AST viewer: for each function definition, creates a graph of
 // the AST and displays it with the graph viewer "dotty".  Also outputs
 // function declarations to stderr.
 std::unique_ptr<ASTConsumer> CreateASTViewer();
 
 } // end clang namespace
 
 #endif

clang/lib/AST/DeclPrinter.cpp

 void DeclPrinter::printDeclType(QualType T, StringRef DeclName, bool Pack) {
   // Normally, a PackExpansionType is written as T[3]... (for instance, as a
   // template argument), but if it is the type of a declaration, the ellipsis
   // is placed before the name being declared.
   if (auto *PET = T->getAs<PackExpansionType>()) {
     Pack = true;
     T = PET->getPattern();
   }
+  if (Policy.SuppressDeclNames)
+    DeclName = "";
   T.print(Out, Policy, (Pack ? "..." : "") + DeclName, Indentation);
 }
 
 void DeclPrinter::ProcessDeclGroup(SmallVectorImpl<Decl*>& Decls) {
   this->Indent();
   Decl::printGroup(Decls.data(), Decls.size(), Out, Policy, Indentation);
   Out << ";\n";
   Decls.clear();
for (DeclContext::decl_iterator D = DC->decls_begin(), DEnd = DC->decls_end();
     // containing ObjCInterfaceDecl.
     if (isa<ObjCIvarDecl>(*D))
       continue;
 
     // Skip over implicit declarations in pretty-printing mode.
     if (D->isImplicit())
       continue;
 
+    if (isa<EnumDecl>(*D) && Policy.SuppressEnumDecls)
+      continue;
+
+    if (isa<TypedefDecl>(*D) && Policy.SuppressTypedefDecls)
+      continue;
+
+    if (isa<RecordDecl>(*D) && Policy.SuppressRecordDecls)
+      continue;
+
     // Don't print implicit specializations, as they are printed when visiting
     // corresponding templates.
     if (auto FD = dyn_cast<FunctionDecl>(*D))
       if (FD->getTemplateSpecializationKind() == TSK_ImplicitInstantiation &&
           !isa<ClassTemplateSpecializationDecl>(DC))
         continue;
 
     // The next bits of code handle stuff like "struct {int x;} a,b"; we're

clang/lib/Frontend/ASTConsumers.cpp

namespace {
   class ASTPrinter : public ASTConsumer,
                      public RecursiveASTVisitor<ASTPrinter> {
     typedef RecursiveASTVisitor<ASTPrinter> base;
 
   public:
     enum Kind { DumpFull, Dump, Print, None };
     ASTPrinter(std::unique_ptr<raw_ostream> Out, Kind K,
                ASTDumpOutputFormat Format, StringRef FilterString,
-               bool DumpLookups = false, bool DumpDeclTypes = false)
+               StringRef FilterNodeTypeString, bool DumpLookups = false,
+               bool DumpDeclTypes = false)
         : Out(Out ? *Out : llvm::outs()), OwnedOut(std::move(Out)),
           OutputKind(K), OutputFormat(Format), FilterString(FilterString),
-          DumpLookups(DumpLookups), DumpDeclTypes(DumpDeclTypes) {}
+          FilterNodeTypeString(FilterNodeTypeString), DumpLookups(DumpLookups),
+          DumpDeclTypes(DumpDeclTypes) {}
 
     void HandleTranslationUnit(ASTContext &Context) override {
       TranslationUnitDecl *D = Context.getTranslationUnitDecl();
 
       if (FilterString.empty())
         return print(D);
 
       TraverseDecl(D);
private:
     std::string getName(Decl *D) {
       if (isa<NamedDecl>(D))
         return cast<NamedDecl>(D)->getQualifiedNameAsString();
       return "";
     }
     bool filterMatches(Decl *D) {
       return getName(D).find(FilterString) != std::string::npos;
     }
+
+    void adjustPrintingPolicy(PrintingPolicy &Policy) {
+      if (FilterNodeTypeString.find("record") != std::string::npos)
+        Policy.SuppressRecordDecls = true;
+      if (FilterNodeTypeString.find("typedef") != std::string::npos)
+        Policy.SuppressTypedefDecls = true;
+      if (FilterNodeTypeString.find("enum") != std::string::npos)
+        Policy.SuppressEnumDecls = true;
+      if (FilterNodeTypeString.find("declnames") != std::string::npos)
+        Policy.SuppressDeclNames = true;
+
+      Policy.PolishForDeclaration = true;
+    }
+
     void print(Decl *D) {
       if (DumpLookups) {
         if (DeclContext *DC = dyn_cast<DeclContext>(D)) {
           if (DC == DC->getPrimaryContext())
             DC->dumpLookups(Out, OutputKind != None, OutputKind == DumpFull);
           else
             Out << "Lookup map is in primary DeclContext "
                 << DC->getPrimaryContext() << "\n";
         } else
           Out << "Not a DeclContext\n";
       } else if (OutputKind == Print) {
         PrintingPolicy Policy(D->getASTContext().getLangOpts());
+        adjustPrintingPolicy(Policy);
         D->print(Out, Policy, /*Indentation=*/0, /*PrintInstantiation=*/true);
       } else if (OutputKind != None) {
         D->dump(Out, OutputKind == DumpFull, OutputFormat);
       }
 
       if (DumpDeclTypes) {
         Decl *InnerD = D;
         if (auto *TD = dyn_cast<TemplateDecl>(D))
private:
     Kind OutputKind;
 
     /// What format should the output take?
     ASTDumpOutputFormat OutputFormat;
 
     /// Which declarations or DeclContexts to display.
     std::string FilterString;
 
+    /// Ignore certain node types
+    std::string FilterNodeTypeString;
+
     /// Whether the primary output is lookup results or declarations. Individual
     /// results will be output with a format determined by OutputKind. This is
     /// incompatible with OutputKind == Print.
     bool DumpLookups;
 
     /// Whether to dump the type for each declaration dumped.
     bool DumpDeclTypes;
   };
namespace {
 
   private:
     raw_ostream &Out;
   };
 } // end anonymous namespace
 
 std::unique_ptr<ASTConsumer>
 clang::CreateASTPrinter(std::unique_ptr<raw_ostream> Out,
-                        StringRef FilterString) {
+                        StringRef FilterString,
+                        StringRef FilterNodeTypeString) {
   return std::make_unique<ASTPrinter>(std::move(Out), ASTPrinter::Print,
-                                       ADOF_Default, FilterString);
+                                      ADOF_Default, FilterString,
+                                      FilterNodeTypeString);
 }
 
 std::unique_ptr<ASTConsumer>
 clang::CreateASTDumper(std::unique_ptr<raw_ostream> Out, StringRef FilterString,
                        bool DumpDecls, bool Deserialize, bool DumpLookups,
-                       bool DumpDeclTypes, ASTDumpOutputFormat Format) {
+                       bool DumpDeclTypes, ASTDumpOutputFormat Format,
+                       StringRef FilterNodeTypeString) {
   assert((DumpDecls || Deserialize || DumpLookups) && "nothing to dump");
   return std::make_unique<ASTPrinter>(
       std::move(Out),
       Deserialize ? ASTPrinter::DumpFull
                   : DumpDecls ? ASTPrinter::Dump : ASTPrinter::None,
-      Format, FilterString, DumpLookups, DumpDeclTypes);
+      Format, FilterString, FilterNodeTypeString, DumpLookups, DumpDeclTypes);
 }
 
 std::unique_ptr<ASTConsumer> clang::CreateASTDeclNodeLister() {
   return std::make_unique<ASTDeclNodeLister>(nullptr);
 }
 
 //===----------------------------------------------------------------------===//
 /// ASTViewer - AST Visualization

clang/test/SemaOpenCL/compare-header-and-tablegen.cl

+// RUN: clang-tblgen -gen-clang-opencl-builtin-header %S/../../lib/Sema/OpenCLBuiltins.td > %t1
+// RUN: sort %t1  -o %t1.sorted
+// RUN: clang-check -extra-arg=-cl-std=CL2.0 --ast-print --ast-dump-node-type-filter=record,typedef,enum,declnames %S/../../lib/Headers/opencl-c.h  > %t2
+// RUN: sort %t2  -o %t2.sorted
+// RUN: diff %t1.sorted %t2.sorted

clang/tools/clang-check/ClangCheck.cpp

 static cl::opt<bool>
     ASTPrint("ast-print",
              cl::desc(Options.getOptionHelpText(options::OPT_ast_print)),
              cl::cat(ClangCheckCategory));
 static cl::opt<std::string> ASTDumpFilter(
     "ast-dump-filter",
     cl::desc(Options.getOptionHelpText(options::OPT_ast_dump_filter)),
     cl::cat(ClangCheckCategory));
+static cl::opt<std::string>
+    ASTDumpNodeTypeFilter("ast-dump-node-type-filter",
+                          cl::desc("Ignore some of the node types"),
+                          cl::cat(ClangCheckCategory));
 static cl::opt<bool>
     Analyze("analyze",
             cl::desc(Options.getOptionHelpText(options::OPT_analyze)),
             cl::cat(ClangCheckCategory));
 
 static cl::opt<bool>
     Fixit("fixit", cl::desc(Options.getOptionHelpText(options::OPT_fixit)),
           cl::cat(ClangCheckCategory));
std::unique_ptr<clang::ASTConsumer> newASTConsumer() {
     if (ASTDump)
       return clang::CreateASTDumper(nullptr /*Dump to stdout.*/, ASTDumpFilter,
                                     /*DumpDecls=*/true,
                                     /*Deserialize=*/false,
                                     /*DumpLookups=*/false,
                                     /*DumpDeclTypes=*/false,
                                     clang::ADOF_Default);
     if (ASTPrint)
-      return clang::CreateASTPrinter(nullptr, ASTDumpFilter);
+      return clang::CreateASTPrinter(nullptr, ASTDumpFilter,
+                                     ASTDumpNodeTypeFilter);
     return std::make_unique<clang::ASTConsumer>();
   }
 };
 
 } // namespace
 
 int main(int argc, const char **argv) {
   llvm::sys::PrintStackTraceOnErrorSignal(argv[0]);

clang/utils/TableGen/ClangOpenCLBuiltinEmitter.cpp

private:
   // Then we will create a mapping of the vector of signatures:
   // SignatureListMap[<I1, I2, I3>] = <
   //                  <"cos", "sin">,
   //                  <Signature1, Signature2, Signature3>>
   // The function "tan", having the same signatures, would be mapped to the
   // same entry (<I1, I2, I3>).
   MapVector<BuiltinIndexListTy *, BuiltinTableEntries> SignatureListMap;
 };
+
+// OpenCL builtin header emitter.  This class processes the same TableGen input
+// as BuiltinNameEmitter, but generates a header-like output of all the
+// declarations which are contained in TableGen file
+class OpenCLBuiltinHeaderEmitter {
+public:
+  OpenCLBuiltinHeaderEmitter(RecordKeeper &Records, raw_ostream &OS)
+      : Records(Records), OS(OS) {}
+
+  // Entrypoint to generate the header
+  void Emit();
+
+private:
+  struct TypeFlags {
+    TypeFlags() : IsConst(false), IsVolatile(false), IsPointer(false) {}
+    bool IsConst : 1;
+    bool IsVolatile : 1;
+    bool IsPointer : 1;
+    StringRef AddrSpace;
+  };
+
+  // Return a string representation of the given type, such that it can be
+  // used as a type in OpenCL C code.
+  std::string getTypeString(const Record *Type, TypeFlags Flags,
+                            int VectorSize) const;
+
+  // Return the type(s) and vector size(s) for the given type.  For
+  // non-GenericTypes, the resulting vectors will contain 1 element.  For
+  // GenericTypes, the resulting vectors typically contain multiple elements.
+  void getTypeLists(Record *Type, TypeFlags &Flags,
+                    std::vector<Record *> &TypeList,
+                    std::vector<int64_t> &VectorList) const;
+
+  // Expand the TableGen Records representing a builtin function signature into
+  // one or more function signatures.  Return them as a vector of a vector of
+  // strings, with each string containing an OpenCL C type and optional
+  // qualifiers.
+  //
+  // The Records may contain GenericTypes, which expand into multiple
+  // signatures.  Repeated occurrences of GenericType in a signature expand to
+  // the same types.  For example [char, FGenType, FGenType] expands to:
+  //   [char, float, float]
+  //   [char, float2, float2]
+  //   [char, float3, float3]
+  //   ...
+  void
+  expandTypesInSignature(const std::vector<Record *> &Signature,
+                         SmallVectorImpl<SmallVector<std::string, 2>> &Types);
+
+  // Contains OpenCL builtin functions and related information, stored as
+  // Record instances. They are coming from the associated TableGen file.
+  RecordKeeper &Records;
+
+  // The output file.
+  raw_ostream &OS;
+};
+
 } // namespace
 
 void BuiltinNameEmitter::Emit() {
   emitSourceFileHeader("OpenCL Builtin handling", OS);
 
   OS << "#include \"llvm/ADT/StringRef.h\"\n";
   OS << "using namespace clang;\n\n";
 
if (Ty.IsPointer != 0) {
     }
   }
 )";
 
   // End of the "OCL2Qual" function.
   OS << "\n} // OCL2Qual\n";
 }
 
+std::string OpenCLBuiltinHeaderEmitter::getTypeString(const Record *Type,
+                                                      TypeFlags Flags,
+                                                      int VectorSize) const {
+  std::string S;
+  if (Type->getValueAsBit("IsConst") || Flags.IsConst) {
+    S += "const ";
+  }
+  if (Type->getValueAsBit("IsVolatile") || Flags.IsVolatile) {
+    S += "volatile ";
+  }
+
+  auto PrintAddrSpace = [&S](StringRef AddrSpace) {
+    S += StringSwitch<const char *>(AddrSpace)
+             .Case("clang::LangAS::opencl_private", "__private")
+             .Case("clang::LangAS::opencl_global", "__global")
+             .Case("clang::LangAS::opencl_constant", "__constant")
+             .Case("clang::LangAS::opencl_local", "__local")
+             .Case("clang::LangAS::opencl_generic", "__generic")
+             .Default("__private");
+    S += " ";
+  };
+  if (Flags.IsPointer) {
+    PrintAddrSpace(Flags.AddrSpace);
+  } else if (Type->getValueAsBit("IsPointer")) {
+    PrintAddrSpace(Type->getValueAsString("AddrSpace"));
+  }
+
+  StringRef Acc = Type->getValueAsString("AccessQualifier");
+  if (Acc != "") {
+    S += StringSwitch<const char *>(Acc)
+             .Case("RO", "__read_only ")
+             .Case("WO", "__write_only ")
+             .Case("RW", "__read_write ");
+  }
+
+  S += Type->getValueAsString("Name").str();
+  if (VectorSize > 1) {
+    S += std::to_string(VectorSize);
+  }
+
+  if (Type->getValueAsBit("IsPointer") || Flags.IsPointer) {
+    S += " *";
+  }
+
+  return S;
+}
+
+void OpenCLBuiltinHeaderEmitter::getTypeLists(
+    Record *Type, TypeFlags &Flags, std::vector<Record *> &TypeList,
+    std::vector<int64_t> &VectorList) const {
+  bool isGenType = Type->isSubClassOf("GenericType");
+  if (isGenType) {
+    TypeList = Type->getValueAsDef("TypeList")->getValueAsListOfDefs("List");
+    VectorList =
+        Type->getValueAsDef("VectorList")->getValueAsListOfInts("List");
+    return;
+  } else if (Type->isSubClassOf("PointerType") ||
+             Type->isSubClassOf("ConstType") ||
+             Type->isSubClassOf("VolatileType")) {
+    StringRef SubTypeName = Type->getValueAsString("Name");
+    Record *PossibleGenType = Records.getDef(SubTypeName);
+    if (PossibleGenType && PossibleGenType->isSubClassOf("GenericType")) {
+      // When PointerType, ConstType, or VolatileType is applied to a
+      // GenericType, the flags need to be taken from the subtype, not from the
+      // GenericType.
+      Flags.IsPointer = Type->getValueAsBit("IsPointer");
+      Flags.IsConst = Type->getValueAsBit("IsConst");
+      Flags.IsVolatile = Type->getValueAsBit("IsVolatile");
+      Flags.AddrSpace = Type->getValueAsString("AddrSpace");
+      getTypeLists(PossibleGenType, Flags, TypeList, VectorList);
+      return;
+    }
+  }
+
+  // Not a GenericType, so just insert the single type.
+  TypeList.push_back(Type);
+  VectorList.push_back(Type->getValueAsInt("VecWidth"));
+}
+
+void OpenCLBuiltinHeaderEmitter::expandTypesInSignature(
+    const std::vector<Record *> &Signature,
+    SmallVectorImpl<SmallVector<std::string, 2>> &Types) {
+  // Find out if there are any GenTypes in this signature, and if so, calculate
+  // into how many signatures they will expand.
+  unsigned NumSignatures = 1;
+  SmallVector<SmallVector<std::string, 4>, 4> ExpandedGenTypes;
+  for (const auto &Arg : Signature) {
+    SmallVector<std::string, 4> ExpandedArg;
+    std::vector<Record *> TypeList;
+    std::vector<int64_t> VectorList;
+    TypeFlags Flags;
+
+    getTypeLists(Arg, Flags, TypeList, VectorList);
+
+    // Insert the Cartesian product of the types and vector sizes.
+    for (const auto &Vector : VectorList) {
+      for (const auto &Type : TypeList) {
+        ExpandedArg.push_back(getTypeString(Type, Flags, Vector));
+      }
+    }
+    NumSignatures = std::max<unsigned>(NumSignatures, ExpandedArg.size());
+    ExpandedGenTypes.push_back(ExpandedArg);
+  }
+
+  // Now the total number of signatures is known.  Populate the return list with
+  // all signatures.
+  for (unsigned I = 0; I < NumSignatures; I++) {
+    SmallVector<std::string, 2> Args;
+
+    // Process a single signature.
+    for (unsigned ArgNum = 0; ArgNum < Signature.size(); ArgNum++) {
+      // For differently-sized GenTypes in a parameter list, the smaller
+      // GenTypes just repeat.
+      Args.push_back(
+          ExpandedGenTypes[ArgNum][I % ExpandedGenTypes[ArgNum].size()]);
+    }
+    Types.push_back(Args);
+  }
+}
+
+void OpenCLBuiltinHeaderEmitter::Emit() {
+  // Iterate over all builtins.
+  std::vector<Record *> Builtins = Records.getAllDerivedDefinitions("Builtin");
+
+  for (const auto *B : Builtins) {
+    StringRef Name = B->getValueAsString("Name");
+
+    SmallVector<SmallVector<std::string, 2>, 4> FTypes;
+    expandTypesInSignature(B->getValueAsListOfDefs("Signature"), FTypes);
+
+    for (const auto &Signature : FTypes) {
+      // Emit function declaration.
+      OS << Signature[0] << ' ';
+      OS << Name;
+      OS << '(';
+      if (Signature.size() > 1) {
+        for (unsigned I = 1; I < Signature.size(); I++) {
+          if (I != 1)
+            OS << ", ";
+          OS << Signature[I];
+        }
+      }
+      OS << ");\n";
+    }
+  }
+}
+
 void clang::EmitClangOpenCLBuiltins(RecordKeeper &Records, raw_ostream &OS) {
   BuiltinNameEmitter NameChecker(Records, OS);
   NameChecker.Emit();
 }
+
+void clang::EmitClangOpenCLBuiltinHeader(RecordKeeper &Records,
+                                         raw_ostream &OS) {
+  OpenCLBuiltinHeaderEmitter HeaderGenerator(Records, OS);
+  HeaderGenerator.Emit();
+}

clang/utils/TableGen/TableGen.cpp

enum ActionType {
   GenClangSyntaxNodeList,
   GenClangSyntaxNodeClasses,
   GenClangCommentHTMLTags,
   GenClangCommentHTMLTagsProperties,
   GenClangCommentHTMLNamedCharacterReferences,
   GenClangCommentCommandInfo,
   GenClangCommentCommandList,
   GenClangOpenCLBuiltins,
+  GenClangOpenCLBuiltinHeader,
   GenArmNeon,
   GenArmFP16,
   GenArmBF16,
   GenArmNeonSema,
   GenArmNeonTest,
   GenArmMveHeader,
   GenArmMveBuiltinDef,
   GenArmMveBuiltinSema,
cl::values(
         clEnumValN(GenClangCommentCommandInfo, "gen-clang-comment-command-info",
                    "Generate command properties for commands that "
                    "are used in documentation comments"),
         clEnumValN(GenClangCommentCommandList, "gen-clang-comment-command-list",
                    "Generate list of commands that are used in "
                    "documentation comments"),
         clEnumValN(GenClangOpenCLBuiltins, "gen-clang-opencl-builtins",
                    "Generate OpenCL builtin declaration handlers"),
+        clEnumValN(GenClangOpenCLBuiltinHeader,
+                   "gen-clang-opencl-builtin-header",
+                   "Generate OpenCL builtin header from TableGen file"),
         clEnumValN(GenArmNeon, "gen-arm-neon", "Generate arm_neon.h for clang"),
         clEnumValN(GenArmFP16, "gen-arm-fp16", "Generate arm_fp16.h for clang"),
         clEnumValN(GenArmBF16, "gen-arm-bf16", "Generate arm_bf16.h for clang"),
         clEnumValN(GenArmNeonSema, "gen-arm-neon-sema",
                    "Generate ARM NEON sema support for clang"),
         clEnumValN(GenArmNeonTest, "gen-arm-neon-test",
                    "Generate ARM NEON tests for clang"),
         clEnumValN(GenArmSveHeader, "gen-arm-sve-header",
case GenClangCommentCommandInfo:
     EmitClangCommentCommandInfo(Records, OS);
     break;
   case GenClangCommentCommandList:
     EmitClangCommentCommandList(Records, OS);
     break;
   case GenClangOpenCLBuiltins:
     EmitClangOpenCLBuiltins(Records, OS);
     break;
+  case GenClangOpenCLBuiltinHeader:
+    EmitClangOpenCLBuiltinHeader(Records, OS);
+    break;
   case GenClangSyntaxNodeList:
     EmitClangSyntaxNodeList(Records, OS);
     break;
   case GenClangSyntaxNodeClasses:
     EmitClangSyntaxNodeClasses(Records, OS);
     break;
   case GenArmNeon:
     EmitNeon(Records, OS);

clang/utils/TableGen/TableGenBackends.h

 void EmitCdeBuiltinAliases(llvm::RecordKeeper &Records, llvm::raw_ostream &OS);
 
 void EmitClangAttrDocs(llvm::RecordKeeper &Records, llvm::raw_ostream &OS);
 void EmitClangDiagDocs(llvm::RecordKeeper &Records, llvm::raw_ostream &OS);
 void EmitClangOptDocs(llvm::RecordKeeper &Records, llvm::raw_ostream &OS);
 
 void EmitClangOpenCLBuiltins(llvm::RecordKeeper &Records,
                              llvm::raw_ostream &OS);
+void EmitClangOpenCLBuiltinHeader(llvm::RecordKeeper &Records,
+                                  llvm::raw_ostream &OS);
 
 void EmitClangDataCollectors(llvm::RecordKeeper &Records,
                              llvm::raw_ostream &OS);
 
 void EmitTestPragmaAttributeSupportedAttributes(llvm::RecordKeeper &Records,
                                                 llvm::raw_ostream &OS);
 
 } // end namespace clang
 
 #endif