svenvh created this revision.
svenvh added reviewers: azabaznov, Anastasia.
svenvh added a project: clang.
Herald added subscribers: ldrumm, yaxunl.
svenvh requested review of this revision.
Herald added a subscriber: cfe-commits.
Add an emitter to produce something similar to `opencl-c.h` from the
OpenCL builtin descriptions in `OpenCLBuiltins.td`
Factor out functionality that can be shared between the test emitter
and the header emitter into a new base class.
This only adds the emitter, without any use of it currently.
Example output excerpt (the full output is about 16k lines):
...
__ovld __cnfn size_t get_global_id(uint);
__ovld __cnfn size_t get_local_size(uint);
__ovld __cnfn size_t get_local_id(uint);
__ovld __cnfn size_t get_num_groups(uint);
__ovld __cnfn size_t get_group_id(uint);
__ovld __cnfn size_t get_global_offset(uint);
#if __OPENCL_C_VERSION__ >= CL_VERSION_2_0
__ovld size_t get_enqueued_local_size(uint);
#endif // MinVersion
...
Repository:
rG LLVM Github Monorepo
https://reviews.llvm.org/D104040
Files:
clang/utils/TableGen/ClangOpenCLBuiltinEmitter.cpp
clang/utils/TableGen/TableGen.cpp
clang/utils/TableGen/TableGenBackends.h
Index: clang/utils/TableGen/TableGenBackends.h
===================================================================
--- clang/utils/TableGen/TableGenBackends.h
+++ clang/utils/TableGen/TableGenBackends.h
@@ -122,6 +122,8 @@
void EmitClangOpenCLBuiltins(llvm::RecordKeeper &Records,
llvm::raw_ostream &OS);
+void EmitClangOpenCLBuiltinHeader(llvm::RecordKeeper &Records,
+ llvm::raw_ostream &OS);
void EmitClangOpenCLBuiltinTests(llvm::RecordKeeper &Records,
llvm::raw_ostream &OS);
Index: clang/utils/TableGen/TableGen.cpp
===================================================================
--- clang/utils/TableGen/TableGen.cpp
+++ clang/utils/TableGen/TableGen.cpp
@@ -63,6 +63,7 @@
GenClangCommentCommandInfo,
GenClangCommentCommandList,
GenClangOpenCLBuiltins,
+ GenClangOpenCLBuiltinHeader,
GenClangOpenCLBuiltinTests,
GenArmNeon,
GenArmFP16,
@@ -195,6 +196,9 @@
"documentation comments"),
clEnumValN(GenClangOpenCLBuiltins, "gen-clang-opencl-builtins",
"Generate OpenCL builtin declaration handlers"),
+ clEnumValN(GenClangOpenCLBuiltinHeader,
+ "gen-clang-opencl-builtin-header",
+ "Generate OpenCL builtin header"),
clEnumValN(GenClangOpenCLBuiltinTests, "gen-clang-opencl-builtin-tests",
"Generate OpenCL builtin declaration tests"),
clEnumValN(GenArmNeon, "gen-arm-neon", "Generate arm_neon.h for clang"),
@@ -374,6 +378,9 @@
case GenClangOpenCLBuiltins:
EmitClangOpenCLBuiltins(Records, OS);
break;
+ case GenClangOpenCLBuiltinHeader:
+ EmitClangOpenCLBuiltinHeader(Records, OS);
+ break;
case GenClangOpenCLBuiltinTests:
EmitClangOpenCLBuiltinTests(Records, OS);
break;
Index: clang/utils/TableGen/ClangOpenCLBuiltinEmitter.cpp
===================================================================
--- clang/utils/TableGen/ClangOpenCLBuiltinEmitter.cpp
+++ clang/utils/TableGen/ClangOpenCLBuiltinEmitter.cpp
@@ -229,19 +229,17 @@
MapVector<BuiltinIndexListTy *, BuiltinTableEntries> SignatureListMap;
};
-// OpenCL builtin test generator. This class processes the same TableGen input
-// as BuiltinNameEmitter, but generates a .cl file that contains a call to each
-// builtin function described in the .td input.
-class OpenCLBuiltinTestEmitter {
+/// Base class for emitting a file (e.g. header or test) from OpenCLBuiltins.td
+class OpenCLBuiltinFileEmitterBase {
public:
- OpenCLBuiltinTestEmitter(RecordKeeper &Records, raw_ostream &OS)
+ OpenCLBuiltinFileEmitterBase(RecordKeeper &Records, raw_ostream &OS)
: Records(Records), OS(OS) {}
// Entrypoint to generate the functions for testing all OpenCL builtin
// functions.
- void emit();
+ virtual void emit() = 0;
-private:
+protected:
struct TypeFlags {
TypeFlags() : IsConst(false), IsVolatile(false), IsPointer(false) {}
bool IsConst : 1;
@@ -278,6 +276,18 @@
expandTypesInSignature(const std::vector<Record *> &Signature,
SmallVectorImpl<SmallVector<std::string, 2>> &Types);
+ // Emit extension enabling pragmas.
+ void emitExtensionSetup();
+
+ // Emit an #if guard for a Builtin's extension. Return the corresponding
+ // closing #endif, or an empty string if no extension #if guard was emitted.
+ std::string emitExtensionGuard(const Record *Builtin);
+
+ // Emit an #if guard for a Builtin's language version. Return the
+ // corresponding closing #endif, or an empty string if no version #if guard
+ // was emitted.
+ std::string emitVersionGuard(const Record *Builtin);
+
// Contains OpenCL builtin functions and related information, stored as
// Record instances. They are coming from the associated TableGen file.
RecordKeeper &Records;
@@ -286,6 +296,31 @@
raw_ostream &OS;
};
+// OpenCL builtin test generator. This class processes the same TableGen input
+// as BuiltinNameEmitter, but generates a .cl file that contains a call to each
+// builtin function described in the .td input.
+class OpenCLBuiltinTestEmitter : public OpenCLBuiltinFileEmitterBase {
+public:
+ OpenCLBuiltinTestEmitter(RecordKeeper &Records, raw_ostream &OS)
+ : OpenCLBuiltinFileEmitterBase(Records, OS) {}
+
+ // Entrypoint to generate the functions for testing all OpenCL builtin
+ // functions.
+ void emit() override;
+};
+
+// OpenCL builtin header generator. This class processes the same TableGen
+// input as BuiltinNameEmitter, but generates a .h file that contains a
+// prototype for each builtin function described in the .td input.
+class OpenCLBuiltinHeaderEmitter : public OpenCLBuiltinFileEmitterBase {
+public:
+ OpenCLBuiltinHeaderEmitter(RecordKeeper &Records, raw_ostream &OS)
+ : OpenCLBuiltinFileEmitterBase(Records, OS) {}
+
+ // Entrypoint to generate the header.
+ void emit() override;
+};
+
} // namespace
void BuiltinNameEmitter::Emit() {
@@ -919,9 +954,9 @@
OS << "\n} // OCL2Qual\n";
}
-std::string OpenCLBuiltinTestEmitter::getTypeString(const Record *Type,
- TypeFlags Flags,
- int VectorSize) const {
+std::string OpenCLBuiltinFileEmitterBase::getTypeString(const Record *Type,
+ TypeFlags Flags,
+ int VectorSize) const {
std::string S;
if (Type->getValueAsBit("IsConst") || Flags.IsConst) {
S += "const ";
@@ -966,7 +1001,7 @@
return S;
}
-void OpenCLBuiltinTestEmitter::getTypeLists(
+void OpenCLBuiltinFileEmitterBase::getTypeLists(
Record *Type, TypeFlags &Flags, std::vector<Record *> &TypeList,
std::vector<int64_t> &VectorList) const {
bool isGenType = Type->isSubClassOf("GenericType");
@@ -999,7 +1034,7 @@
VectorList.push_back(Type->getValueAsInt("VecWidth"));
}
-void OpenCLBuiltinTestEmitter::expandTypesInSignature(
+void OpenCLBuiltinFileEmitterBase::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
@@ -1040,10 +1075,7 @@
}
}
-void OpenCLBuiltinTestEmitter::emit() {
- emitSourceFileHeader("OpenCL Builtin exhaustive testing", OS);
-
- // Enable some extensions for testing.
+void OpenCLBuiltinFileEmitterBase::emitExtensionSetup() {
OS << R"(
#pragma OPENCL EXTENSION cl_khr_fp16 : enable
#pragma OPENCL EXTENSION cl_khr_fp64 : enable
@@ -1054,6 +1086,60 @@
#pragma OPENCL EXTENSION cl_khr_3d_image_writes : enable
)";
+}
+
+std::string
+OpenCLBuiltinFileEmitterBase::emitExtensionGuard(const Record *Builtin) {
+ StringRef Extensions =
+ Builtin->getValueAsDef("Extension")->getValueAsString("ExtName");
+ if (Extensions.empty())
+ return "";
+
+ OS << "#if";
+
+ SmallVector<StringRef, 2> ExtVec;
+ Extensions.split(ExtVec, " ");
+ bool isFirst = true;
+ for (StringRef Ext : ExtVec) {
+ if (!isFirst) {
+ OS << " &&";
+ }
+ OS << " defined(" << Ext << ")";
+ isFirst = false;
+ }
+ OS << "\n";
+
+ return "#endif // Extension\n";
+}
+
+std::string
+OpenCLBuiltinFileEmitterBase::emitVersionGuard(const Record *Builtin) {
+ std::string OptionalEndif;
+ auto PrintOpenCLVersion = [this](int Version) {
+ OS << "CL_VERSION_" << (Version / 100) << "_" << ((Version % 100) / 10);
+ };
+ int MinVersion = Builtin->getValueAsDef("MinVersion")->getValueAsInt("ID");
+ if (MinVersion != 100) {
+ // OpenCL 1.0 is the default minimum version.
+ OS << "#if __OPENCL_C_VERSION__ >= ";
+ PrintOpenCLVersion(MinVersion);
+ OS << "\n";
+ OptionalEndif = "#endif // MinVersion\n" + OptionalEndif;
+ }
+ int MaxVersion = Builtin->getValueAsDef("MaxVersion")->getValueAsInt("ID");
+ if (MaxVersion) {
+ OS << "#if __OPENCL_C_VERSION__ < ";
+ PrintOpenCLVersion(MaxVersion);
+ OS << "\n";
+ OptionalEndif = "#endif // MaxVersion\n" + OptionalEndif;
+ }
+ return OptionalEndif;
+}
+
+void OpenCLBuiltinTestEmitter::emit() {
+ emitSourceFileHeader("OpenCL Builtin exhaustive testing", OS);
+
+ emitExtensionSetup();
// Ensure each test has a unique name by numbering them.
unsigned TestID = 0;
@@ -1067,43 +1153,10 @@
expandTypesInSignature(B->getValueAsListOfDefs("Signature"), FTypes);
OS << "// Test " << Name << "\n";
- std::string OptionalEndif;
- StringRef Extensions =
- B->getValueAsDef("Extension")->getValueAsString("ExtName");
- if (!Extensions.empty()) {
- OS << "#if";
- OptionalEndif = "#endif // Extension\n";
-
- SmallVector<StringRef, 2> ExtVec;
- Extensions.split(ExtVec, " ");
- bool isFirst = true;
- for (StringRef Ext : ExtVec) {
- if (!isFirst) {
- OS << " &&";
- }
- OS << " defined(" << Ext << ")";
- isFirst = false;
- }
- OS << "\n";
- }
- auto PrintOpenCLVersion = [this](int Version) {
- OS << "CL_VERSION_" << (Version / 100) << "_" << ((Version % 100) / 10);
- };
- int MinVersion = B->getValueAsDef("MinVersion")->getValueAsInt("ID");
- if (MinVersion != 100) {
- // OpenCL 1.0 is the default minimum version.
- OS << "#if __OPENCL_C_VERSION__ >= ";
- PrintOpenCLVersion(MinVersion);
- OS << "\n";
- OptionalEndif = "#endif // MinVersion\n" + OptionalEndif;
- }
- int MaxVersion = B->getValueAsDef("MaxVersion")->getValueAsInt("ID");
- if (MaxVersion) {
- OS << "#if __OPENCL_C_VERSION__ < ";
- PrintOpenCLVersion(MaxVersion);
- OS << "\n";
- OptionalEndif = "#endif // MaxVersion\n" + OptionalEndif;
- }
+
+ std::string OptionalExtensionEndif = emitExtensionGuard(B);
+ std::string OptionalVersionEndif = emitVersionGuard(B);
+
for (const auto &Signature : FTypes) {
// Emit function declaration.
OS << Signature[0] << " test" << TestID++ << "_" << Name << "(";
@@ -1132,15 +1185,76 @@
// End of function body.
OS << "}\n";
}
- OS << OptionalEndif << "\n";
+
+ OS << OptionalVersionEndif;
+ OS << OptionalExtensionEndif;
}
}
+void OpenCLBuiltinHeaderEmitter::emit() {
+ emitSourceFileHeader("OpenCL Builtin declarations", OS);
+
+ emitExtensionSetup();
+
+ OS << R"(
+#define __ovld __attribute__((overloadable))
+#define __conv __attribute__((convergent))
+#define __purefn __attribute__((pure))
+#define __cnfn __attribute__((const))
+
+)";
+
+ // Iterate over all builtins.
+ std::vector<Record *> Builtins = Records.getAllDerivedDefinitions("Builtin");
+ for (const auto *B : Builtins) {
+ StringRef Name = B->getValueAsString("Name");
+
+ std::string OptionalExtensionEndif = emitExtensionGuard(B);
+ std::string OptionalVersionEndif = emitVersionGuard(B);
+
+ SmallVector<SmallVector<std::string, 2>, 4> FTypes;
+ expandTypesInSignature(B->getValueAsListOfDefs("Signature"), FTypes);
+
+ for (const auto &Signature : FTypes) {
+ // Emit function declaration.
+ OS << "__ovld ";
+ if (B->getValueAsBit("IsConst"))
+ OS << "__cnfn ";
+ if (B->getValueAsBit("IsPure"))
+ OS << "__purefn ";
+ if (B->getValueAsBit("IsConv"))
+ OS << "__conv ";
+
+ OS << Signature[0] << " " << Name << "(";
+ if (Signature.size() > 1) {
+ for (unsigned I = 1; I < Signature.size(); I++) {
+ if (I != 1)
+ OS << ", ";
+ OS << Signature[I];
+ }
+ }
+ OS << ");\n";
+ }
+
+ OS << OptionalVersionEndif;
+ OS << OptionalExtensionEndif;
+ }
+
+ OS << "\n// Disable any extensions we may have enabled previously.\n"
+ "#pragma OPENCL EXTENSION all : disable";
+}
+
void clang::EmitClangOpenCLBuiltins(RecordKeeper &Records, raw_ostream &OS) {
BuiltinNameEmitter NameChecker(Records, OS);
NameChecker.Emit();
}
+void clang::EmitClangOpenCLBuiltinHeader(RecordKeeper &Records,
+ raw_ostream &OS) {
+ OpenCLBuiltinHeaderEmitter HeaderFileGenerator(Records, OS);
+ HeaderFileGenerator.emit();
+}
+
void clang::EmitClangOpenCLBuiltinTests(RecordKeeper &Records,
raw_ostream &OS) {
OpenCLBuiltinTestEmitter TestFileGenerator(Records, OS);
_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
