[clang] 231aa30 - [OpenACC][CIR] Basic infrastructure for OpenACC lowering (#134717)
via cfe-commits
cfe-commits at lists.llvm.org
Tue Apr 8 10:06:31 PDT 2025
Author: Erich Keane
Date: 2025-04-08T10:06:28-07:00
New Revision: 231aa3070dcd91e10e9972d20f7557c0068c41e3
URL: https://github.com/llvm/llvm-project/commit/231aa3070dcd91e10e9972d20f7557c0068c41e3
DIFF: https://github.com/llvm/llvm-project/commit/231aa3070dcd91e10e9972d20f7557c0068c41e3.diff
LOG: [OpenACC][CIR] Basic infrastructure for OpenACC lowering (#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.
Added:
clang/lib/CIR/CodeGen/CIRGenDeclOpenACC.cpp
clang/lib/CIR/CodeGen/CIRGenStmtOpenACC.cpp
clang/test/CIR/CodeGenOpenACC/openacc-not-implemented-global.cpp
clang/test/CIR/CodeGenOpenACC/openacc-not-implemented.cpp
clang/test/Driver/openacc-no-cir.c
Modified:
clang/include/clang/AST/DeclOpenACC.h
clang/include/clang/AST/GlobalDecl.h
clang/include/clang/Basic/DiagnosticDriverKinds.td
clang/lib/AST/DeclOpenACC.cpp
clang/lib/CIR/CodeGen/CIRGenDecl.cpp
clang/lib/CIR/CodeGen/CIRGenFunction.h
clang/lib/CIR/CodeGen/CIRGenModule.cpp
clang/lib/CIR/CodeGen/CIRGenModule.h
clang/lib/CIR/CodeGen/CIRGenStmt.cpp
clang/lib/CIR/CodeGen/CMakeLists.txt
clang/lib/Frontend/CompilerInvocation.cpp
Removed:
################################################################################
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 df11a79a56b3b..baf5371d2682d 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 a96b4fb33390c..c69ad3adc5b3b 100644
--- a/clang/include/clang/Basic/DiagnosticDriverKinds.td
+++ b/clang/include/clang/Basic/DiagnosticDriverKinds.td
@@ -850,4 +850,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/CIRGenDeclOpenACC.cpp b/clang/lib/CIR/CodeGen/CIRGenDeclOpenACC.cpp
new file mode 100644
index 0000000000000..b588a50aa0404
--- /dev/null
+++ b/clang/lib/CIR/CodeGen/CIRGenDeclOpenACC.cpp
@@ -0,0 +1,34 @@
+//===----------------------------------------------------------------------===//
+//
+// 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");
+ else
+ llvm_unreachable("unknown OpenACC declaration kind?");
+}
diff --git a/clang/lib/CIR/CodeGen/CIRGenFunction.h b/clang/lib/CIR/CodeGen/CIRGenFunction.h
index 1bedbe28ae625..f505ed8fff311 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..f0e9b0349f709 100644
--- a/clang/lib/CIR/CodeGen/CIRGenModule.cpp
+++ b/clang/lib/CIR/CodeGen/CIRGenModule.cpp
@@ -16,6 +16,7 @@
#include "clang/AST/ASTContext.h"
#include "clang/AST/DeclBase.h"
+#include "clang/AST/DeclOpenACC.h"
#include "clang/AST/GlobalDecl.h"
#include "clang/Basic/SourceManager.h"
#include "clang/CIR/Dialect/IR/CIRDialect.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/CIRGenStmt.cpp b/clang/lib/CIR/CodeGen/CIRGenStmt.cpp
index 00d33e7feddff..2551a670b5325 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;
@@ -85,7 +86,34 @@ mlir::LogicalResult CIRGenFunction::emitStmt(const Stmt *s,
return emitWhileStmt(cast<WhileStmt>(*s));
case Stmt::DoStmtClass:
return emitDoStmt(cast<DoStmt>(*s));
-
+ 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:
+ return emitOpenACCAtomicConstruct(cast<OpenACCAtomicConstruct>(*s));
case Stmt::OMPScopeDirectiveClass:
case Stmt::OMPErrorDirectiveClass:
case Stmt::NoStmtClass:
@@ -192,20 +220,6 @@ mlir::LogicalResult CIRGenFunction::emitStmt(const Stmt *s,
case Stmt::OMPAssumeDirectiveClass:
case Stmt::OMPMaskedDirectiveClass:
case Stmt::OMPStripeDirectiveClass:
- case Stmt::OpenACCComputeConstructClass:
- case Stmt::OpenACCLoopConstructClass:
- case Stmt::OpenACCCombinedConstructClass:
- case Stmt::OpenACCDataConstructClass:
- case Stmt::OpenACCEnterDataConstructClass:
- case Stmt::OpenACCExitDataConstructClass:
- case Stmt::OpenACCHostDataConstructClass:
- case Stmt::OpenACCWaitConstructClass:
- case Stmt::OpenACCInitConstructClass:
- case Stmt::OpenACCShutdownConstructClass:
- case Stmt::OpenACCSetConstructClass:
- case Stmt::OpenACCUpdateConstructClass:
- case Stmt::OpenACCCacheConstructClass:
- case Stmt::OpenACCAtomicConstructClass:
case Stmt::ObjCAtCatchStmtClass:
case Stmt::ObjCAtFinallyStmtClass:
cgm.errorNYI(s->getSourceRange(),
diff --git a/clang/lib/CIR/CodeGen/CIRGenStmtOpenACC.cpp b/clang/lib/CIR/CodeGen/CIRGenStmtOpenACC.cpp
new file mode 100644
index 0000000000000..cbae170162ffe
--- /dev/null
+++ b/clang/lib/CIR/CodeGen/CIRGenStmtOpenACC.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/CMakeLists.txt b/clang/lib/CIR/CodeGen/CMakeLists.txt
index da8d63ca569af..11902c708c505 100644
--- a/clang/lib/CIR/CodeGen/CMakeLists.txt
+++ b/clang/lib/CIR/CodeGen/CMakeLists.txt
@@ -9,6 +9,7 @@ get_property(dialect_libs GLOBAL PROPERTY MLIR_DIALECT_LIBS)
add_clang_library(clangCIR
CIRGenerator.cpp
CIRGenDecl.cpp
+ CIRGenDeclOpenACC.cpp
CIRGenExpr.cpp
CIRGenExprAggregate.cpp
CIRGenExprConstant.cpp
@@ -16,6 +17,7 @@ add_clang_library(clangCIR
CIRGenFunction.cpp
CIRGenModule.cpp
CIRGenStmt.cpp
+ CIRGenStmtOpenACC.cpp
CIRGenTypes.cpp
DEPENDS
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 at +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..61bed79dc14ea
--- /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 at +2{{ClangIR code gen Not Yet Implemented: OpenACC Compute Construct}}
+// expected-error at +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 at +2{{ClangIR code gen Not Yet Implemented: OpenACC Loop Construct}}
+// expected-error at +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 at +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..7b67df2b6b886
--- /dev/null
+++ b/clang/test/Driver/openacc-no-cir.c
@@ -0,0 +1,6 @@
+// RUN: %clang -fopenacc -S %s 2>&1 | FileCheck %s -check-prefix=ERROR
+// RUN: %clang -fclangir -fopenacc -S %s 2>&1 | FileCheck %s --allow-empty -check-prefix=NOERROR
+// RUN: %clang -fopenacc -fclangir -S %s 2>&1 | FileCheck %s --allow-empty -check-prefix=NOERROR
+
+// ERROR: OpenACC directives will result in no runtime behavior; use -fclangir to enable runtime effect
+// NOERROR-NOT: OpenACC directives
More information about the cfe-commits
mailing list