[OpenACC][CIR] Basic infrastructure for OpenACC lowering #134717
Merged
Conversation
This file contains hidden or bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
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.
llvmbot
added
clang
|
@llvm/pr-subscribers-clang @llvm/pr-subscribers-clangir Author: Erich Keane (erichkeane) ChangesThis 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. Patch is 20.72 KiB, truncated to 20.00 KiB below, full version: https://github.com/llvm/llvm-project/pull/134717.diff 16 Files Affected:
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: O...
[truncated]
|
|
✅ With the latest revision this PR passed the C/C++ code formatter. |
bcardosolopes
approved these changes
Apr 7, 2025
| 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}} |
There was a problem hiding this comment.
super nit: the indentation is odd, might look better with @-1 and @-2 (or alternatively be indented with the # below?
There was a problem hiding this comment.
sigh... yeah, #pragma ends up getting indented automatically to be left-aligned, not scope-aligned, which is a touch annoying (Clang-format AND vim do it). So I've moved the comments to the 'left' so at least they are aligned.
andykaylor
approved these changes
Apr 8, 2025
| errorNYI(d->getSourceRange(), "OpenACC Routine Construct"); | ||
| else if (isa<OpenACCDeclareDecl>(d)) | ||
| errorNYI(d->getSourceRange(), "OpenACC Declare Construct"); | ||
| } |
There was a problem hiding this comment.
Should there be an errorNYI or llvm_unreachable call for the case where this is neither of the above?
There was a problem hiding this comment.
yeah, unreachable. I'll add it.
| //===--------------------------------------------------------------------===// | ||
| public: | ||
| mlir::LogicalResult | ||
| emitOpenACCComputeConstruct(const OpenACCComputeConstruct &S); |
There was a problem hiding this comment.
Suggested change
| emitOpenACCComputeConstruct(const OpenACCComputeConstruct &S); | |
| emitOpenACCComputeConstruct(const OpenACCComputeConstruct &s); |
MLIR naming style throughout the changes to this header
There was a problem hiding this comment.
Ooof... I was doing so well :D I caught myself on this one a dozen times so far, I'm guessing beating the old coding standard out of my fingers is going to take a while :)
clang/lib/CIR/CodeGen/CIRGenStmt.cpp
Outdated
| case Stmt::OpenACCComputeConstructClass: | ||
| return emitOpenACCComputeConstruct(cast<OpenACCComputeConstruct>(*s)); |
There was a problem hiding this comment.
Maybe move these above the NYI cases?
clang/test/Driver/openacc-no-cir.c
Outdated
| // 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 |
There was a problem hiding this comment.
Can you use -Werror instead of the negative check?
There was a problem hiding this comment.
Unfortunately not, driver warnings are... special and don't really pay attention to Werror since they happen "before" Werror has been processed.
I HAVE however modified the test run-lines in a way that are perhaps better, in that they don't try to link now, so hopefully they do a better job diagnosing. I've also made the 'NOT' line a little less specific, so hopefully it'll not go stale as easy.
| @@ -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 " | |||
There was a problem hiding this comment.
Suggested change
| : Warning<"OpenACC directives will result in no runtime behavior, use " | |
| : Warning<"OpenACC directives will result in no runtime behavior. Use " |
There was a problem hiding this comment.
We typically don't do full-stops during diagnostics as they are not sentences. we sometimes use semi-colons though, which is probably more fitting here.
|
I should be all caught up on review comments, but am going to be done for the evening. Let me know if theres anything else I should do! |
|
LLVM Buildbot has detected a new failure on builder Full details are available at: https://lab.llvm.org/buildbot/#/builders/141/builds/7716 Here is the relevant piece of the build log for the reference |
pranavk
added a commit
that referenced
this pull request
Apr 10, 2025
We execute tests in read only environment which leads to test failure when tests try to write to the current directory. Either they should write to a temporary directory or not write if output is not needed. Fallback from #134717
var-const
pushed a commit
to ldionne/llvm-project
that referenced
this pull request
Apr 17, 2025
…35242) We execute tests in read only environment which leads to test failure when tests try to write to the current directory. Either they should write to a temporary directory or not write if output is not needed. Fallback from llvm#134717
Sign up for free
to join this conversation on GitHub.
Already have an account?
Sign in to comment
Add this suggestion to a batch that can be applied as a single commit.
This suggestion is invalid because no changes were made to the code.
Suggestions cannot be applied while the pull request is closed.
Suggestions cannot be applied while viewing a subset of changes.
Only one suggestion per line can be applied in a batch.
Add this suggestion to a batch that can be applied as a single commit.
Applying suggestions on deleted lines is not supported.
You must change the existing code in this line in order to create a valid suggestion.
Outdated suggestions cannot be applied.
This suggestion has been applied or marked resolved.
Suggestions cannot be applied from pending reviews.
Suggestions cannot be applied on multi-line comments.
Suggestions cannot be applied while the pull request is queued to merge.
Suggestion cannot be applied right now. Please check back later.
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.