https://github.com/erichkeane created
https://github.com/llvm/llvm-project/pull/134717
This is the first of a few patches that will do infrastructure work to enable
the OpenACC lowering via the OpenACC dialect.
At the moment this just gets the various function calls that will end up
generating OpenACC, plus some tests to validate that we're doing the
diagnostics in OpenACC specific locations.
Additionally, this adds Stmt and Decl files for CIRGen.
>From 0c0892e5142e2d0f687c091cb55ea14c97c975eb Mon Sep 17 00:00:00 2001
From: erichkeane <eke...@nvidia.com>
Date: Mon, 7 Apr 2025 08:07:18 -0700
Subject: [PATCH] [OpenACC][CIR] Basic infrastructure for OpenACC lowering
This is the first of a few patches that will do infrastructure work to
enable the OpenACC lowering via the OpenACC dialect.
At the moment this just gets the various function calls that will end up
generating OpenACC, plus some tests to validate that we're doing the
diagnostics in OpenACC specific locations.
Additionally, this adds Stmt and Decl files for CIRGen.
---
clang/include/clang/AST/DeclOpenACC.h | 2 +
clang/include/clang/AST/GlobalDecl.h | 3 +
.../clang/Basic/DiagnosticDriverKinds.td | 5 +
clang/lib/AST/DeclOpenACC.cpp | 5 +
clang/lib/CIR/CodeGen/CIRGenDecl.cpp | 7 ++
clang/lib/CIR/CodeGen/CIRGenFunction.h | 30 ++++++
clang/lib/CIR/CodeGen/CIRGenModule.cpp | 12 +++
clang/lib/CIR/CodeGen/CIRGenModule.h | 2 +
clang/lib/CIR/CodeGen/CIRGenOpenACCDecl.cpp | 33 +++++++
clang/lib/CIR/CodeGen/CIRGenOpenACCStmt.cpp | 91 +++++++++++++++++++
clang/lib/CIR/CodeGen/CIRGenStmt.cpp | 25 ++++-
clang/lib/CIR/CodeGen/CMakeLists.txt | 2 +
clang/lib/Frontend/CompilerInvocation.cpp | 49 ++++++++++
.../openacc-not-implemented-global.cpp | 6 ++
.../openacc-not-implemented.cpp | 20 ++++
clang/test/Driver/openacc-no-cir.c | 6 ++
16 files changed, 293 insertions(+), 5 deletions(-)
create mode 100644 clang/lib/CIR/CodeGen/CIRGenOpenACCDecl.cpp
create mode 100644 clang/lib/CIR/CodeGen/CIRGenOpenACCStmt.cpp
create mode 100644
clang/test/CIR/CodeGenOpenACC/openacc-not-implemented-global.cpp
create mode 100644 clang/test/CIR/CodeGenOpenACC/openacc-not-implemented.cpp
create mode 100644 clang/test/Driver/openacc-no-cir.c
diff --git a/clang/include/clang/AST/DeclOpenACC.h
b/clang/include/clang/AST/DeclOpenACC.h
index 26cf721561fb1..8c612fbf1ec07 100644
--- a/clang/include/clang/AST/DeclOpenACC.h
+++ b/clang/include/clang/AST/DeclOpenACC.h
@@ -59,6 +59,8 @@ class OpenACCConstructDecl : public Decl {
}
ArrayRef<const OpenACCClause *> clauses() const { return Clauses; }
+ static bool classof(const Decl *D) { return classofKind(D->getKind()); }
+ static bool classofKind(Kind K);
};
class OpenACCDeclareDecl final
diff --git a/clang/include/clang/AST/GlobalDecl.h
b/clang/include/clang/AST/GlobalDecl.h
index 386693cabb1fb..8bf2eb2a5b78f 100644
--- a/clang/include/clang/AST/GlobalDecl.h
+++ b/clang/include/clang/AST/GlobalDecl.h
@@ -17,6 +17,7 @@
#include "clang/AST/Attr.h"
#include "clang/AST/DeclCXX.h"
#include "clang/AST/DeclObjC.h"
+#include "clang/AST/DeclOpenACC.h"
#include "clang/AST/DeclOpenMP.h"
#include "clang/AST/DeclTemplate.h"
#include "clang/Basic/ABI.h"
@@ -86,6 +87,8 @@ class GlobalDecl {
GlobalDecl(const ObjCMethodDecl *D) { Init(D); }
GlobalDecl(const OMPDeclareReductionDecl *D) { Init(D); }
GlobalDecl(const OMPDeclareMapperDecl *D) { Init(D); }
+ GlobalDecl(const OpenACCRoutineDecl *D) { Init(D); }
+ GlobalDecl(const OpenACCDeclareDecl *D) { Init(D); }
GlobalDecl(const CXXConstructorDecl *D, CXXCtorType Type) : Value(D, Type) {}
GlobalDecl(const CXXDestructorDecl *D, CXXDtorType Type) : Value(D, Type) {}
GlobalDecl(const VarDecl *D, DynamicInitKind StubKind)
diff --git a/clang/include/clang/Basic/DiagnosticDriverKinds.td
b/clang/include/clang/Basic/DiagnosticDriverKinds.td
index df24cca49aaae..3f671c88f3148 100644
--- a/clang/include/clang/Basic/DiagnosticDriverKinds.td
+++ b/clang/include/clang/Basic/DiagnosticDriverKinds.td
@@ -849,4 +849,9 @@ def warn_missing_include_dirs : Warning<
def err_drv_malformed_warning_suppression_mapping : Error<
"failed to process suppression mapping file '%0': %1">;
+
+def warn_drv_openacc_without_cir
+ : Warning<"OpenACC directives will result in no runtime behavior, use "
+ "-fclangir to enable runtime effect">,
+ InGroup<SourceUsesOpenACC>;
}
diff --git a/clang/lib/AST/DeclOpenACC.cpp b/clang/lib/AST/DeclOpenACC.cpp
index 760c08d21cccd..e0fe7be8fc1a3 100644
--- a/clang/lib/AST/DeclOpenACC.cpp
+++ b/clang/lib/AST/DeclOpenACC.cpp
@@ -17,6 +17,11 @@
using namespace clang;
+bool OpenACCConstructDecl::classofKind(Kind K) {
+ return OpenACCDeclareDecl::classofKind(K) ||
+ OpenACCRoutineDecl::classofKind(K);
+}
+
OpenACCDeclareDecl *
OpenACCDeclareDecl::Create(ASTContext &Ctx, DeclContext *DC,
SourceLocation StartLoc, SourceLocation DirLoc,
diff --git a/clang/lib/CIR/CodeGen/CIRGenDecl.cpp
b/clang/lib/CIR/CodeGen/CIRGenDecl.cpp
index 5b832b463e752..d0eb648683e8c 100644
--- a/clang/lib/CIR/CodeGen/CIRGenDecl.cpp
+++ b/clang/lib/CIR/CodeGen/CIRGenDecl.cpp
@@ -15,6 +15,7 @@
#include "mlir/IR/Location.h"
#include "clang/AST/Attr.h"
#include "clang/AST/Decl.h"
+#include "clang/AST/DeclOpenACC.h"
#include "clang/AST/Expr.h"
#include "clang/AST/ExprCXX.h"
#include "clang/CIR/MissingFeatures.h"
@@ -266,6 +267,12 @@ void CIRGenFunction::emitDecl(const Decl &d) {
emitVarDecl(vd);
return;
}
+ case Decl::OpenACCDeclare:
+ emitOpenACCDeclare(cast<OpenACCDeclareDecl>(d));
+ return;
+ case Decl::OpenACCRoutine:
+ emitOpenACCRoutine(cast<OpenACCRoutineDecl>(d));
+ return;
default:
cgm.errorNYI(d.getSourceRange(), "emitDecl: unhandled decl type");
}
diff --git a/clang/lib/CIR/CodeGen/CIRGenFunction.h
b/clang/lib/CIR/CodeGen/CIRGenFunction.h
index 1bedbe28ae625..c498ca04b90f8 100644
--- a/clang/lib/CIR/CodeGen/CIRGenFunction.h
+++ b/clang/lib/CIR/CodeGen/CIRGenFunction.h
@@ -509,6 +509,36 @@ class CIRGenFunction : public CIRGenTypeCache {
public:
Address createTempAlloca(mlir::Type ty, CharUnits align, mlir::Location loc,
const Twine &name, bool insertIntoFnEntryBlock);
+
+
//===--------------------------------------------------------------------===//
+ // OpenACC Emission
+
//===--------------------------------------------------------------------===//
+public:
+ mlir::LogicalResult
+ emitOpenACCComputeConstruct(const OpenACCComputeConstruct &S);
+ mlir::LogicalResult emitOpenACCLoopConstruct(const OpenACCLoopConstruct &S);
+ mlir::LogicalResult
+ emitOpenACCCombinedConstruct(const OpenACCCombinedConstruct &S);
+ mlir::LogicalResult emitOpenACCDataConstruct(const OpenACCDataConstruct &S);
+ mlir::LogicalResult
+ emitOpenACCEnterDataConstruct(const OpenACCEnterDataConstruct &S);
+ mlir::LogicalResult
+ emitOpenACCExitDataConstruct(const OpenACCExitDataConstruct &S);
+ mlir::LogicalResult
+ emitOpenACCHostDataConstruct(const OpenACCHostDataConstruct &S);
+ mlir::LogicalResult emitOpenACCWaitConstruct(const OpenACCWaitConstruct &S);
+ mlir::LogicalResult emitOpenACCInitConstruct(const OpenACCInitConstruct &S);
+ mlir::LogicalResult
+ emitOpenACCShutdownConstruct(const OpenACCShutdownConstruct &S);
+ mlir::LogicalResult emitOpenACCSetConstruct(const OpenACCSetConstruct &S);
+ mlir::LogicalResult
+ emitOpenACCUpdateConstruct(const OpenACCUpdateConstruct &S);
+ mlir::LogicalResult
+ emitOpenACCAtomicConstruct(const OpenACCAtomicConstruct &S);
+ mlir::LogicalResult emitOpenACCCacheConstruct(const OpenACCCacheConstruct
&S);
+
+ void emitOpenACCDeclare(const OpenACCDeclareDecl &D);
+ void emitOpenACCRoutine(const OpenACCRoutineDecl &D);
};
} // namespace clang::CIRGen
diff --git a/clang/lib/CIR/CodeGen/CIRGenModule.cpp
b/clang/lib/CIR/CodeGen/CIRGenModule.cpp
index d3b3b0632c2f0..e88de98a927c8 100644
--- a/clang/lib/CIR/CodeGen/CIRGenModule.cpp
+++ b/clang/lib/CIR/CodeGen/CIRGenModule.cpp
@@ -17,6 +17,7 @@
#include "clang/AST/ASTContext.h"
#include "clang/AST/DeclBase.h"
#include "clang/AST/GlobalDecl.h"
+#include "clang/AST/DeclOpenACC.h"
#include "clang/Basic/SourceManager.h"
#include "clang/CIR/Dialect/IR/CIRDialect.h"
#include "clang/CIR/MissingFeatures.h"
@@ -91,6 +92,11 @@ mlir::Location CIRGenModule::getLoc(SourceRange cRange) {
}
void CIRGenModule::emitGlobal(clang::GlobalDecl gd) {
+ if (const auto *cd = dyn_cast<clang::OpenACCConstructDecl>(gd.getDecl())) {
+ emitGlobalOpenACCDecl(cd);
+ return;
+ }
+
const auto *global = cast<ValueDecl>(gd.getDecl());
if (const auto *fd = dyn_cast<FunctionDecl>(global)) {
@@ -423,6 +429,12 @@ void CIRGenModule::emitTopLevelDecl(Decl *decl) {
emitGlobal(vd);
break;
}
+ case Decl::OpenACCRoutine:
+ emitGlobalOpenACCDecl(cast<OpenACCRoutineDecl>(decl));
+ break;
+ case Decl::OpenACCDeclare:
+ emitGlobalOpenACCDecl(cast<OpenACCDeclareDecl>(decl));
+ break;
}
}
diff --git a/clang/lib/CIR/CodeGen/CIRGenModule.h
b/clang/lib/CIR/CodeGen/CIRGenModule.h
index 6ba1ccc4ddd9f..ab4545effde45 100644
--- a/clang/lib/CIR/CodeGen/CIRGenModule.h
+++ b/clang/lib/CIR/CodeGen/CIRGenModule.h
@@ -113,6 +113,8 @@ class CIRGenModule : public CIRGenTypeCache {
void emitGlobalVarDefinition(const clang::VarDecl *vd,
bool isTentative = false);
+ void emitGlobalOpenACCDecl(const clang::OpenACCConstructDecl *cd);
+
/// Return the result of value-initializing the given type, i.e. a null
/// expression of the given type.
mlir::Value emitNullConstant(QualType t, mlir::Location loc);
diff --git a/clang/lib/CIR/CodeGen/CIRGenOpenACCDecl.cpp
b/clang/lib/CIR/CodeGen/CIRGenOpenACCDecl.cpp
new file mode 100644
index 0000000000000..488e51e09666b
--- /dev/null
+++ b/clang/lib/CIR/CodeGen/CIRGenOpenACCDecl.cpp
@@ -0,0 +1,33 @@
+//===----------------------------------------------------------------------===//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM
Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+//
+//===----------------------------------------------------------------------===//
+//
+// This contains code to emit Decl nodes as CIR code.
+//
+//===----------------------------------------------------------------------===//
+
+#include "CIRGenFunction.h"
+#include "clang/AST/DeclOpenACC.h"
+
+using namespace clang;
+using namespace clang::CIRGen;
+
+void CIRGenFunction::emitOpenACCDeclare(const OpenACCDeclareDecl &d) {
+ getCIRGenModule().errorNYI(d.getSourceRange(), "OpenACC Declare Construct");
+}
+
+void CIRGenFunction::emitOpenACCRoutine(const OpenACCRoutineDecl &d) {
+ getCIRGenModule().errorNYI(d.getSourceRange(), "OpenACC Routine Construct");
+}
+
+void CIRGenModule::emitGlobalOpenACCDecl(const OpenACCConstructDecl *d) {
+ if (isa<OpenACCRoutineDecl>(d))
+ errorNYI(d->getSourceRange(), "OpenACC Routine Construct");
+ else if (isa<OpenACCDeclareDecl>(d))
+ errorNYI(d->getSourceRange(), "OpenACC Declare Construct");
+
+}
diff --git a/clang/lib/CIR/CodeGen/CIRGenOpenACCStmt.cpp
b/clang/lib/CIR/CodeGen/CIRGenOpenACCStmt.cpp
new file mode 100644
index 0000000000000..cbae170162ffe
--- /dev/null
+++ b/clang/lib/CIR/CodeGen/CIRGenOpenACCStmt.cpp
@@ -0,0 +1,91 @@
+//===----------------------------------------------------------------------===//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM
Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+//
+//===----------------------------------------------------------------------===//
+//
+// Emit OpenACC Stmt nodes as CIR code.
+//
+//===----------------------------------------------------------------------===//
+
+#include "CIRGenBuilder.h"
+#include "CIRGenFunction.h"
+#include "clang/AST/StmtOpenACC.h"
+
+using namespace clang;
+using namespace clang::CIRGen;
+using namespace cir;
+
+mlir::LogicalResult
+CIRGenFunction::emitOpenACCComputeConstruct(const OpenACCComputeConstruct &s) {
+ getCIRGenModule().errorNYI(s.getSourceRange(), "OpenACC Compute Construct");
+ return mlir::failure();
+}
+
+mlir::LogicalResult
+CIRGenFunction::emitOpenACCLoopConstruct(const OpenACCLoopConstruct &s) {
+ getCIRGenModule().errorNYI(s.getSourceRange(), "OpenACC Loop Construct");
+ return mlir::failure();
+}
+mlir::LogicalResult CIRGenFunction::emitOpenACCCombinedConstruct(
+ const OpenACCCombinedConstruct &s) {
+ getCIRGenModule().errorNYI(s.getSourceRange(), "OpenACC Combined Construct");
+ return mlir::failure();
+}
+mlir::LogicalResult
+CIRGenFunction::emitOpenACCDataConstruct(const OpenACCDataConstruct &s) {
+ getCIRGenModule().errorNYI(s.getSourceRange(), "OpenACC Data Construct");
+ return mlir::failure();
+}
+mlir::LogicalResult CIRGenFunction::emitOpenACCEnterDataConstruct(
+ const OpenACCEnterDataConstruct &s) {
+ getCIRGenModule().errorNYI(s.getSourceRange(), "OpenACC EnterData
Construct");
+ return mlir::failure();
+}
+mlir::LogicalResult CIRGenFunction::emitOpenACCExitDataConstruct(
+ const OpenACCExitDataConstruct &s) {
+ getCIRGenModule().errorNYI(s.getSourceRange(), "OpenACC ExitData Construct");
+ return mlir::failure();
+}
+mlir::LogicalResult CIRGenFunction::emitOpenACCHostDataConstruct(
+ const OpenACCHostDataConstruct &s) {
+ getCIRGenModule().errorNYI(s.getSourceRange(), "OpenACC HostData Construct");
+ return mlir::failure();
+}
+mlir::LogicalResult
+CIRGenFunction::emitOpenACCWaitConstruct(const OpenACCWaitConstruct &s) {
+ getCIRGenModule().errorNYI(s.getSourceRange(), "OpenACC Wait Construct");
+ return mlir::failure();
+}
+mlir::LogicalResult
+CIRGenFunction::emitOpenACCInitConstruct(const OpenACCInitConstruct &s) {
+ getCIRGenModule().errorNYI(s.getSourceRange(), "OpenACC Init Construct");
+ return mlir::failure();
+}
+mlir::LogicalResult CIRGenFunction::emitOpenACCShutdownConstruct(
+ const OpenACCShutdownConstruct &s) {
+ getCIRGenModule().errorNYI(s.getSourceRange(), "OpenACC Shutdown Construct");
+ return mlir::failure();
+}
+mlir::LogicalResult
+CIRGenFunction::emitOpenACCSetConstruct(const OpenACCSetConstruct &s) {
+ getCIRGenModule().errorNYI(s.getSourceRange(), "OpenACC Set Construct");
+ return mlir::failure();
+}
+mlir::LogicalResult
+CIRGenFunction::emitOpenACCUpdateConstruct(const OpenACCUpdateConstruct &s) {
+ getCIRGenModule().errorNYI(s.getSourceRange(), "OpenACC Update Construct");
+ return mlir::failure();
+}
+mlir::LogicalResult
+CIRGenFunction::emitOpenACCAtomicConstruct(const OpenACCAtomicConstruct &s) {
+ getCIRGenModule().errorNYI(s.getSourceRange(), "OpenACC Atomic Construct");
+ return mlir::failure();
+}
+mlir::LogicalResult
+CIRGenFunction::emitOpenACCCacheConstruct(const OpenACCCacheConstruct &s) {
+ getCIRGenModule().errorNYI(s.getSourceRange(), "OpenACC Cache Construct");
+ return mlir::failure();
+}
diff --git a/clang/lib/CIR/CodeGen/CIRGenStmt.cpp
b/clang/lib/CIR/CodeGen/CIRGenStmt.cpp
index 00d33e7feddff..4dac9698c518c 100644
--- a/clang/lib/CIR/CodeGen/CIRGenStmt.cpp
+++ b/clang/lib/CIR/CodeGen/CIRGenStmt.cpp
@@ -16,6 +16,7 @@
#include "mlir/IR/Builders.h"
#include "clang/AST/ExprCXX.h"
#include "clang/AST/Stmt.h"
+#include "clang/AST/StmtOpenACC.h"
using namespace clang;
using namespace clang::CIRGen;
@@ -192,25 +193,39 @@ mlir::LogicalResult CIRGenFunction::emitStmt(const Stmt
*s,
case Stmt::OMPAssumeDirectiveClass:
case Stmt::OMPMaskedDirectiveClass:
case Stmt::OMPStripeDirectiveClass:
+ case Stmt::ObjCAtCatchStmtClass:
+ case Stmt::ObjCAtFinallyStmtClass:
+ cgm.errorNYI(s->getSourceRange(),
+ std::string("emitStmt: ") + s->getStmtClassName());
+ return mlir::failure();
case Stmt::OpenACCComputeConstructClass:
+ return emitOpenACCComputeConstruct(cast<OpenACCComputeConstruct>(*s));
case Stmt::OpenACCLoopConstructClass:
+ return emitOpenACCLoopConstruct(cast<OpenACCLoopConstruct>(*s));
case Stmt::OpenACCCombinedConstructClass:
+ return emitOpenACCCombinedConstruct(cast<OpenACCCombinedConstruct>(*s));
case Stmt::OpenACCDataConstructClass:
+ return emitOpenACCDataConstruct(cast<OpenACCDataConstruct>(*s));
case Stmt::OpenACCEnterDataConstructClass:
+ return emitOpenACCEnterDataConstruct(cast<OpenACCEnterDataConstruct>(*s));
case Stmt::OpenACCExitDataConstructClass:
+ return emitOpenACCExitDataConstruct(cast<OpenACCExitDataConstruct>(*s));
case Stmt::OpenACCHostDataConstructClass:
+ return emitOpenACCHostDataConstruct(cast<OpenACCHostDataConstruct>(*s));
case Stmt::OpenACCWaitConstructClass:
+ return emitOpenACCWaitConstruct(cast<OpenACCWaitConstruct>(*s));
case Stmt::OpenACCInitConstructClass:
+ return emitOpenACCInitConstruct(cast<OpenACCInitConstruct>(*s));
case Stmt::OpenACCShutdownConstructClass:
+ return emitOpenACCShutdownConstruct(cast<OpenACCShutdownConstruct>(*s));
case Stmt::OpenACCSetConstructClass:
+ return emitOpenACCSetConstruct(cast<OpenACCSetConstruct>(*s));
case Stmt::OpenACCUpdateConstructClass:
+ return emitOpenACCUpdateConstruct(cast<OpenACCUpdateConstruct>(*s));
case Stmt::OpenACCCacheConstructClass:
+ return emitOpenACCCacheConstruct(cast<OpenACCCacheConstruct>(*s));
case Stmt::OpenACCAtomicConstructClass:
- case Stmt::ObjCAtCatchStmtClass:
- case Stmt::ObjCAtFinallyStmtClass:
- cgm.errorNYI(s->getSourceRange(),
- std::string("emitStmt: ") + s->getStmtClassName());
- return mlir::failure();
+ return emitOpenACCAtomicConstruct(cast<OpenACCAtomicConstruct>(*s));
}
llvm_unreachable("Unexpected statement class");
diff --git a/clang/lib/CIR/CodeGen/CMakeLists.txt
b/clang/lib/CIR/CodeGen/CMakeLists.txt
index da8d63ca569af..51cce54757f99 100644
--- a/clang/lib/CIR/CodeGen/CMakeLists.txt
+++ b/clang/lib/CIR/CodeGen/CMakeLists.txt
@@ -9,12 +9,14 @@ get_property(dialect_libs GLOBAL PROPERTY MLIR_DIALECT_LIBS)
add_clang_library(clangCIR
CIRGenerator.cpp
CIRGenDecl.cpp
+ CIRGenOpenACCDecl.cpp
CIRGenExpr.cpp
CIRGenExprAggregate.cpp
CIRGenExprConstant.cpp
CIRGenExprScalar.cpp
CIRGenFunction.cpp
CIRGenModule.cpp
+ CIRGenOpenACCStmt.cpp
CIRGenStmt.cpp
CIRGenTypes.cpp
diff --git a/clang/lib/Frontend/CompilerInvocation.cpp
b/clang/lib/Frontend/CompilerInvocation.cpp
index 572c71ef1001c..cfc5c069b0849 100644
--- a/clang/lib/Frontend/CompilerInvocation.cpp
+++ b/clang/lib/Frontend/CompilerInvocation.cpp
@@ -4674,6 +4674,51 @@ static bool
isStrictlyPreprocessorAction(frontend::ActionKind Action) {
llvm_unreachable("invalid frontend action");
}
+static bool isCodeGenAction(frontend::ActionKind Action) {
+ switch (Action) {
+ case frontend::EmitAssembly:
+ case frontend::EmitBC:
+ case frontend::EmitCIR:
+ case frontend::EmitHTML:
+ case frontend::EmitLLVM:
+ case frontend::EmitLLVMOnly:
+ case frontend::EmitCodeGenOnly:
+ case frontend::EmitObj:
+ case frontend::GenerateModule:
+ case frontend::GenerateModuleInterface:
+ case frontend::GenerateReducedModuleInterface:
+ case frontend::GenerateHeaderUnit:
+ case frontend::GeneratePCH:
+ case frontend::GenerateInterfaceStubs:
+ return true;
+ case frontend::ASTDeclList:
+ case frontend::ASTDump:
+ case frontend::ASTPrint:
+ case frontend::ASTView:
+ case frontend::ExtractAPI:
+ case frontend::FixIt:
+ case frontend::ParseSyntaxOnly:
+ case frontend::ModuleFileInfo:
+ case frontend::VerifyPCH:
+ case frontend::PluginAction:
+ case frontend::RewriteObjC:
+ case frontend::RewriteTest:
+ case frontend::RunAnalysis:
+ case frontend::TemplightDump:
+ case frontend::DumpCompilerOptions:
+ case frontend::DumpRawTokens:
+ case frontend::DumpTokens:
+ case frontend::InitOnly:
+ case frontend::PrintPreamble:
+ case frontend::PrintPreprocessedInput:
+ case frontend::RewriteMacros:
+ case frontend::RunPreprocessorOnly:
+ case frontend::PrintDependencyDirectivesSourceMinimizerOutput:
+ return false;
+ }
+ llvm_unreachable("invalid frontend action");
+}
+
static void GeneratePreprocessorArgs(const PreprocessorOptions &Opts,
ArgumentConsumer Consumer,
const LangOptions &LangOpts,
@@ -5001,6 +5046,10 @@ bool CompilerInvocation::CreateFromArgsImpl(
Res.getTargetOpts().HostTriple = Res.getFrontendOpts().AuxTriple;
}
+ if (LangOpts.OpenACC && !Res.getFrontendOpts().UseClangIRPipeline &&
+ isCodeGenAction(Res.getFrontendOpts().ProgramAction))
+ Diags.Report(diag::warn_drv_openacc_without_cir);
+
// Set the triple of the host for OpenMP device compile.
if (LangOpts.OpenMPIsTargetDevice)
Res.getTargetOpts().HostTriple = Res.getFrontendOpts().AuxTriple;
diff --git a/clang/test/CIR/CodeGenOpenACC/openacc-not-implemented-global.cpp
b/clang/test/CIR/CodeGenOpenACC/openacc-not-implemented-global.cpp
new file mode 100644
index 0000000000000..2aa32b0484f2c
--- /dev/null
+++ b/clang/test/CIR/CodeGenOpenACC/openacc-not-implemented-global.cpp
@@ -0,0 +1,6 @@
+// RUN: %clang_cc1 -std=c++17 -triple x86_64-unknown-linux-gnu -fopenacc
-fclangir -emit-cir %s -o %t.cir -verify
+// RUN: %clang_cc1 -std=c++17 -triple x86_64-unknown-linux-gnu -fopenacc
-fclangir -emit-llvm %s -o %t-cir.ll -verify
+
+int Global;
+// expected-error@+1{{ClangIR code gen Not Yet Implemented: OpenACC Declare
Construct}}
+#pragma acc declare create(Global)
diff --git a/clang/test/CIR/CodeGenOpenACC/openacc-not-implemented.cpp
b/clang/test/CIR/CodeGenOpenACC/openacc-not-implemented.cpp
new file mode 100644
index 0000000000000..a2fd643c001a3
--- /dev/null
+++ b/clang/test/CIR/CodeGenOpenACC/openacc-not-implemented.cpp
@@ -0,0 +1,20 @@
+// RUN: %clang_cc1 -std=c++17 -triple x86_64-unknown-linux-gnu -fopenacc
-fclangir -emit-cir %s -o %t.cir -verify
+// RUN: %clang_cc1 -std=c++17 -triple x86_64-unknown-linux-gnu -fopenacc
-fclangir -emit-llvm %s -o %t-cir.ll -verify
+
+void HelloWorld(int *A, int *B, int *C, int N) {
+
+ // expected-error@+2{{ClangIR code gen Not Yet Implemented: OpenACC Compute
Construct}}
+ // expected-error@+1{{ClangIR code gen Not Yet Implemented: statement}}
+#pragma acc parallel
+ for (unsigned I = 0; I < N; ++I)
+ A[I] = B[I] + C[I];
+
+ // expected-error@+2{{ClangIR code gen Not Yet Implemented: OpenACC Loop
Construct}}
+ // expected-error@+1{{ClangIR code gen Not Yet Implemented: statement}}
+#pragma acc loop
+ for (unsigned I = 0; I < N; ++I)
+ A[I] = B[I] + C[I];
+
+ // expected-error@+1{{ClangIR code gen Not Yet Implemented: OpenACC Declare
Construct}}
+#pragma acc declare create(A)
+}
diff --git a/clang/test/Driver/openacc-no-cir.c
b/clang/test/Driver/openacc-no-cir.c
new file mode 100644
index 0000000000000..a7f94afbbd889
--- /dev/null
+++ b/clang/test/Driver/openacc-no-cir.c
@@ -0,0 +1,6 @@
+// RUN: not %clang -fopenacc %s 2>&1 | FileCheck %s -check-prefix=ERROR
+// RUN: not %clang -fclangir -fopenacc %s 2>&1 | FileCheck %s
-check-prefix=NOERROR
+// RUN: not %clang -fopenacc -fclangir %s 2>&1 | FileCheck %s
-check-prefix=NOERROR
+
+// ERROR: OpenACC directives will result in no runtime behavior, use -fclangir
to enable runtime effect
+// NOERROR-NOT: OpenACC directives will result in no runtime behavior, use
-fclangir to enable runtime effect
_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
