Author: erichkeane Date: 2025-04-09T11:22:35-07:00 New Revision: 505a7e7d1cd2b77ab67ef82ac4abc9cf94271adb
URL: https://github.com/llvm/llvm-project/commit/505a7e7d1cd2b77ab67ef82ac4abc9cf94271adb DIFF: https://github.com/llvm/llvm-project/commit/505a7e7d1cd2b77ab67ef82ac4abc9cf94271adb.diff LOG: [OpenACC][CIR] Support for `init` and `shutdown` lowering These two are very simple. they don't require any clauses and don't have an associated statement, so they have very simple output. This patch implements them, but none of the associated clauses. Added: clang/test/CIR/CodeGenOpenACC/init.c clang/test/CIR/CodeGenOpenACC/shutdown.c Modified: clang/lib/CIR/CodeGen/CIRGenFunction.h clang/lib/CIR/CodeGen/CIRGenStmtOpenACC.cpp Removed: ################################################################################ diff --git a/clang/lib/CIR/CodeGen/CIRGenFunction.h b/clang/lib/CIR/CodeGen/CIRGenFunction.h index c30fcc2a05f87..b1bfb5594b0db 100644 --- a/clang/lib/CIR/CodeGen/CIRGenFunction.h +++ b/clang/lib/CIR/CodeGen/CIRGenFunction.h @@ -571,6 +571,10 @@ class CIRGenFunction : public CIRGenTypeCache { // OpenACC Emission //===--------------------------------------------------------------------===// private: + template <typename Op> + mlir::LogicalResult + emitOpenACCOp(mlir::Location start, + llvm::ArrayRef<const OpenACCClause *> clauses); // Function to do the basic implementation of an operation with an Associated // Statement. Models AssociatedStmtConstruct. template <typename Op, typename TermOp> diff --git a/clang/lib/CIR/CodeGen/CIRGenStmtOpenACC.cpp b/clang/lib/CIR/CodeGen/CIRGenStmtOpenACC.cpp index e7e56d3602e3a..8c3c87a58c269 100644 --- a/clang/lib/CIR/CodeGen/CIRGenStmtOpenACC.cpp +++ b/clang/lib/CIR/CodeGen/CIRGenStmtOpenACC.cpp @@ -94,6 +94,23 @@ mlir::LogicalResult CIRGenFunction::emitOpenACCOpAssociatedStmt( return res; } +template <typename Op> +mlir::LogicalResult +CIRGenFunction::emitOpenACCOp(mlir::Location start, + llvm::ArrayRef<const OpenACCClause *> clauses) { + mlir::LogicalResult res = mlir::success(); + + llvm::SmallVector<mlir::Type> retTy; + llvm::SmallVector<mlir::Value> operands; + + // Clause-emitter must be here because it might modify operands. + OpenACCClauseCIREmitter clauseEmitter(getCIRGenModule()); + clauseEmitter.VisitClauseList(clauses); + + builder.create<Op>(start, retTy, operands); + return res; +} + mlir::LogicalResult CIRGenFunction::emitOpenACCComputeConstruct(const OpenACCComputeConstruct &s) { mlir::Location start = getLoc(s.getSourceRange().getEnd()); @@ -123,6 +140,17 @@ CIRGenFunction::emitOpenACCDataConstruct(const OpenACCDataConstruct &s) { start, end, s.clauses(), s.getStructuredBlock()); } +mlir::LogicalResult +CIRGenFunction::emitOpenACCInitConstruct(const OpenACCInitConstruct &s) { + mlir::Location start = getLoc(s.getSourceRange().getEnd()); + return emitOpenACCOp<InitOp>(start, s.clauses()); +} +mlir::LogicalResult CIRGenFunction::emitOpenACCShutdownConstruct( + const OpenACCShutdownConstruct &s) { + mlir::Location start = getLoc(s.getSourceRange().getEnd()); + return emitOpenACCOp<ShutdownOp>(start, s.clauses()); +} + mlir::LogicalResult CIRGenFunction::emitOpenACCLoopConstruct(const OpenACCLoopConstruct &s) { getCIRGenModule().errorNYI(s.getSourceRange(), "OpenACC Loop Construct"); @@ -154,16 +182,6 @@ CIRGenFunction::emitOpenACCWaitConstruct(const OpenACCWaitConstruct &s) { 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(); diff --git a/clang/test/CIR/CodeGenOpenACC/init.c b/clang/test/CIR/CodeGenOpenACC/init.c new file mode 100644 index 0000000000000..e81e211b2608f --- /dev/null +++ b/clang/test/CIR/CodeGenOpenACC/init.c @@ -0,0 +1,7 @@ +// RUN: %clang_cc1 -fopenacc -emit-cir -fclangir %s -o - | FileCheck %s + +void acc_init(void) { + // CHECK: cir.func @acc_init() { +#pragma acc init +// CHECK-NEXT: acc.init loc(#{{[a-zA-Z0-9]+}}){{$}} +} diff --git a/clang/test/CIR/CodeGenOpenACC/shutdown.c b/clang/test/CIR/CodeGenOpenACC/shutdown.c new file mode 100644 index 0000000000000..f971807529ecd --- /dev/null +++ b/clang/test/CIR/CodeGenOpenACC/shutdown.c @@ -0,0 +1,7 @@ +// RUN: %clang_cc1 -fopenacc -emit-cir -fclangir %s -o - | FileCheck %s + +void acc_shutdown(void) { + // CHECK: cir.func @acc_shutdown() { +#pragma acc shutdown +// CHECK-NEXT: acc.shutdown loc(#{{[a-zA-Z0-9]+}}){{$}} +} _______________________________________________ cfe-commits mailing list cfe-commits@lists.llvm.org https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
