aboutsummaryrefslogtreecommitdiff
path: root/flang/lib
diff options
context:
space:
mode:
Diffstat (limited to 'flang/lib')
-rw-r--r--flang/lib/Evaluate/intrinsics.cpp4
-rw-r--r--flang/lib/Lower/Bridge.cpp32
-rw-r--r--flang/lib/Lower/IO.cpp4
-rw-r--r--flang/lib/Lower/OpenMP/ClauseProcessor.cpp32
-rw-r--r--flang/lib/Lower/OpenMP/DataSharingProcessor.cpp4
-rw-r--r--flang/lib/Lower/OpenMP/OpenMP.cpp35
-rw-r--r--flang/lib/Lower/OpenMP/Utils.cpp84
-rw-r--r--flang/lib/Lower/OpenMP/Utils.h2
-rw-r--r--flang/lib/Lower/Support/ReductionProcessor.cpp12
-rw-r--r--flang/lib/Optimizer/Builder/IntrinsicCall.cpp64
-rw-r--r--flang/lib/Optimizer/Transforms/ExternalNameConversion.cpp58
-rw-r--r--flang/lib/Semantics/canonicalize-omp.cpp1
12 files changed, 185 insertions, 147 deletions
diff --git a/flang/lib/Evaluate/intrinsics.cpp b/flang/lib/Evaluate/intrinsics.cpp
index 9957010..768e4ba 100644
--- a/flang/lib/Evaluate/intrinsics.cpp
+++ b/flang/lib/Evaluate/intrinsics.cpp
@@ -340,6 +340,7 @@ static const IntrinsicInterface genericIntrinsicFunction[]{
{"acos", {{"x", SameFloating}}, SameFloating},
{"acosd", {{"x", SameFloating}}, SameFloating},
{"acosh", {{"x", SameFloating}}, SameFloating},
+ {"acospi", {{"x", SameFloating}}, SameFloating},
{"adjustl", {{"string", SameChar}}, SameChar},
{"adjustr", {{"string", SameChar}}, SameChar},
{"aimag", {{"z", SameComplex}}, SameReal},
@@ -358,6 +359,7 @@ static const IntrinsicInterface genericIntrinsicFunction[]{
{"asin", {{"x", SameFloating}}, SameFloating},
{"asind", {{"x", SameFloating}}, SameFloating},
{"asinh", {{"x", SameFloating}}, SameFloating},
+ {"asinpi", {{"x", SameFloating}}, SameFloating},
{"associated",
{{"pointer", AnyPointer, Rank::anyOrAssumedRank, Optionality::required,
common::Intent::In, {ArgFlag::canBeNullPointer}},
@@ -957,6 +959,7 @@ static const IntrinsicInterface genericIntrinsicFunction[]{
{"sin", {{"x", SameFloating}}, SameFloating},
{"sind", {{"x", SameFloating}}, SameFloating},
{"sinh", {{"x", SameFloating}}, SameFloating},
+ {"sinpi", {{"x", SameFloating}}, SameFloating},
{"size",
{{"array", AnyData, Rank::arrayOrAssumedRank},
OptionalDIM, // unless array is assumed-size
@@ -988,6 +991,7 @@ static const IntrinsicInterface genericIntrinsicFunction[]{
{"tan", {{"x", SameFloating}}, SameFloating},
{"tand", {{"x", SameFloating}}, SameFloating},
{"tanh", {{"x", SameFloating}}, SameFloating},
+ {"tanpi", {{"x", SameFloating}}, SameFloating},
{"team_number", {OptionalTEAM}, DefaultInt, Rank::scalar,
IntrinsicClass::transformationalFunction},
{"this_image",
diff --git a/flang/lib/Lower/Bridge.cpp b/flang/lib/Lower/Bridge.cpp
index b94833d..92aae79 100644
--- a/flang/lib/Lower/Bridge.cpp
+++ b/flang/lib/Lower/Bridge.cpp
@@ -2125,9 +2125,11 @@ private:
llvm::SmallVector<mlir::Value> reduceVars;
Fortran::lower::omp::ReductionProcessor rp;
- rp.processReductionArguments<fir::DeclareReductionOp>(
+ bool result = rp.processReductionArguments<fir::DeclareReductionOp>(
toLocation(), *this, info.reduceOperatorList, reduceVars,
reduceVarByRef, reductionDeclSymbols, info.reduceSymList);
+ assert(result && "Failed to process `do concurrent` reductions");
+ (void)result;
doConcurrentLoopOp.getReduceVarsMutable().assign(reduceVars);
doConcurrentLoopOp.setReduceSymsAttr(
@@ -5508,10 +5510,34 @@ private:
void genFIR(const Fortran::parser::AssignStmt &stmt) {
const Fortran::semantics::Symbol &symbol =
*std::get<Fortran::parser::Name>(stmt.t).symbol;
+
mlir::Location loc = toLocation();
+ mlir::Type symbolType = genType(symbol);
+ mlir::Value addr = getSymbolAddress(symbol);
+
+ // Handle the case where the assigned variable is declared as a pointer
+ if (auto eleTy = fir::dyn_cast_ptrOrBoxEleTy(symbolType)) {
+ if (auto ptrType = mlir::dyn_cast<fir::PointerType>(eleTy)) {
+ symbolType = ptrType.getEleTy();
+ } else {
+ symbolType = eleTy;
+ }
+ } else if (auto ptrType = mlir::dyn_cast<fir::PointerType>(symbolType)) {
+ symbolType = ptrType.getEleTy();
+ }
+
mlir::Value labelValue = builder->createIntegerConstant(
- loc, genType(symbol), std::get<Fortran::parser::Label>(stmt.t));
- builder->create<fir::StoreOp>(loc, labelValue, getSymbolAddress(symbol));
+ loc, symbolType, std::get<Fortran::parser::Label>(stmt.t));
+
+ // If the address points to a boxed pointer, we need to dereference it
+ if (auto refType = mlir::dyn_cast<fir::ReferenceType>(addr.getType())) {
+ if (auto boxType = mlir::dyn_cast<fir::BoxType>(refType.getEleTy())) {
+ mlir::Value boxValue = builder->create<fir::LoadOp>(loc, addr);
+ addr = builder->create<fir::BoxAddrOp>(loc, boxValue);
+ }
+ }
+
+ builder->create<fir::StoreOp>(loc, labelValue, addr);
}
void genFIR(const Fortran::parser::FormatStmt &) {
diff --git a/flang/lib/Lower/IO.cpp b/flang/lib/Lower/IO.cpp
index c95c340..51f192e 100644
--- a/flang/lib/Lower/IO.cpp
+++ b/flang/lib/Lower/IO.cpp
@@ -468,8 +468,10 @@ getNamelistGroup(Fortran::lower::AbstractConverter &converter,
fir::BoxType boxTy =
fir::BoxType::get(fir::PointerType::get(converter.genType(s)));
auto descFunc = [&](fir::FirOpBuilder &b) {
+ bool couldBeInEquivalence =
+ Fortran::semantics::FindEquivalenceSet(s) != nullptr;
auto box = Fortran::lower::genInitialDataTarget(
- converter, loc, boxTy, *expr, /*couldBeInEquivalence=*/true);
+ converter, loc, boxTy, *expr, couldBeInEquivalence);
fir::HasValueOp::create(b, loc, box);
};
builder.createGlobalConstant(loc, boxTy, mangleName, descFunc, linkOnce);
diff --git a/flang/lib/Lower/OpenMP/ClauseProcessor.cpp b/flang/lib/Lower/OpenMP/ClauseProcessor.cpp
index 8b3ad57..594f95e 100644
--- a/flang/lib/Lower/OpenMP/ClauseProcessor.cpp
+++ b/flang/lib/Lower/OpenMP/ClauseProcessor.cpp
@@ -1116,11 +1116,12 @@ bool ClauseProcessor::processInReduction(
collectReductionSyms(clause, inReductionSyms);
ReductionProcessor rp;
- rp.processReductionArguments<mlir::omp::DeclareReductionOp>(
- currentLocation, converter,
- std::get<typename omp::clause::ReductionOperatorList>(clause.t),
- inReductionVars, inReduceVarByRef, inReductionDeclSymbols,
- inReductionSyms);
+ if (!rp.processReductionArguments<mlir::omp::DeclareReductionOp>(
+ currentLocation, converter,
+ std::get<typename omp::clause::ReductionOperatorList>(clause.t),
+ inReductionVars, inReduceVarByRef, inReductionDeclSymbols,
+ inReductionSyms))
+ inReductionSyms.clear();
// Copy local lists into the output.
llvm::copy(inReductionVars, std::back_inserter(result.inReductionVars));
@@ -1461,10 +1462,12 @@ bool ClauseProcessor::processReduction(
}
ReductionProcessor rp;
- rp.processReductionArguments<mlir::omp::DeclareReductionOp>(
- currentLocation, converter,
- std::get<typename omp::clause::ReductionOperatorList>(clause.t),
- reductionVars, reduceVarByRef, reductionDeclSymbols, reductionSyms);
+ if (!rp.processReductionArguments<mlir::omp::DeclareReductionOp>(
+ currentLocation, converter,
+ std::get<typename omp::clause::ReductionOperatorList>(clause.t),
+ reductionVars, reduceVarByRef, reductionDeclSymbols,
+ reductionSyms))
+ reductionSyms.clear();
// Copy local lists into the output.
llvm::copy(reductionVars, std::back_inserter(result.reductionVars));
llvm::copy(reduceVarByRef, std::back_inserter(result.reductionByref));
@@ -1486,11 +1489,12 @@ bool ClauseProcessor::processTaskReduction(
collectReductionSyms(clause, taskReductionSyms);
ReductionProcessor rp;
- rp.processReductionArguments<mlir::omp::DeclareReductionOp>(
- currentLocation, converter,
- std::get<typename omp::clause::ReductionOperatorList>(clause.t),
- taskReductionVars, taskReduceVarByRef, taskReductionDeclSymbols,
- taskReductionSyms);
+ if (!rp.processReductionArguments<mlir::omp::DeclareReductionOp>(
+ currentLocation, converter,
+ std::get<typename omp::clause::ReductionOperatorList>(clause.t),
+ taskReductionVars, taskReduceVarByRef, taskReductionDeclSymbols,
+ taskReductionSyms))
+ taskReductionSyms.clear();
// Copy local lists into the output.
llvm::copy(taskReductionVars,
std::back_inserter(result.taskReductionVars));
diff --git a/flang/lib/Lower/OpenMP/DataSharingProcessor.cpp b/flang/lib/Lower/OpenMP/DataSharingProcessor.cpp
index 11e4883..2ac4d95 100644
--- a/flang/lib/Lower/OpenMP/DataSharingProcessor.cpp
+++ b/flang/lib/Lower/OpenMP/DataSharingProcessor.cpp
@@ -24,6 +24,7 @@
#include "flang/Optimizer/Dialect/FIROps.h"
#include "flang/Optimizer/HLFIR/HLFIRDialect.h"
#include "flang/Optimizer/HLFIR/HLFIROps.h"
+#include "flang/Parser/openmp-utils.h"
#include "flang/Semantics/attr.h"
#include "flang/Semantics/tools.h"
#include "llvm/ADT/Sequence.h"
@@ -465,7 +466,8 @@ bool DataSharingProcessor::isOpenMPPrivatizingConstruct(
// allow a privatizing clause) are: dispatch, distribute, do, for, loop,
// parallel, scope, sections, simd, single, target, target_data, task,
// taskgroup, taskloop, and teams.
- return llvm::is_contained(privatizing, extractOmpDirective(omp));
+ return llvm::is_contained(privatizing,
+ parser::omp::GetOmpDirectiveName(omp).v);
}
bool DataSharingProcessor::isOpenMPPrivatizingEvaluation(
diff --git a/flang/lib/Lower/OpenMP/OpenMP.cpp b/flang/lib/Lower/OpenMP/OpenMP.cpp
index fc5fef9..4c2d7bad 100644
--- a/flang/lib/Lower/OpenMP/OpenMP.cpp
+++ b/flang/lib/Lower/OpenMP/OpenMP.cpp
@@ -31,6 +31,7 @@
#include "flang/Optimizer/Dialect/FIRType.h"
#include "flang/Optimizer/HLFIR/HLFIROps.h"
#include "flang/Parser/characters.h"
+#include "flang/Parser/openmp-utils.h"
#include "flang/Parser/parse-tree.h"
#include "flang/Semantics/openmp-directive-sets.h"
#include "flang/Semantics/tools.h"
@@ -63,28 +64,6 @@ static void processHostEvalClauses(lower::AbstractConverter &converter,
lower::pft::Evaluation &eval,
mlir::Location loc);
-static llvm::omp::Directive
-getOpenMPDirectiveEnum(const parser::OmpLoopDirective &beginStatment) {
- return beginStatment.v;
-}
-
-static llvm::omp::Directive getOpenMPDirectiveEnum(
- const parser::OmpBeginLoopDirective &beginLoopDirective) {
- return getOpenMPDirectiveEnum(
- std::get<parser::OmpLoopDirective>(beginLoopDirective.t));
-}
-
-static llvm::omp::Directive
-getOpenMPDirectiveEnum(const parser::OpenMPLoopConstruct &ompLoopConstruct) {
- return getOpenMPDirectiveEnum(
- std::get<parser::OmpBeginLoopDirective>(ompLoopConstruct.t));
-}
-
-static llvm::omp::Directive getOpenMPDirectiveEnum(
- const common::Indirection<parser::OpenMPLoopConstruct> &ompLoopConstruct) {
- return getOpenMPDirectiveEnum(ompLoopConstruct.value());
-}
-
namespace {
/// Structure holding information that is needed to pass host-evaluated
/// information to later lowering stages.
@@ -468,7 +447,7 @@ static void processHostEvalClauses(lower::AbstractConverter &converter,
llvm::omp::Directive dir;
auto &nested = parent.getFirstNestedEvaluation();
if (const auto *ompEval = nested.getIf<parser::OpenMPConstruct>())
- dir = extractOmpDirective(*ompEval);
+ dir = parser::omp::GetOmpDirectiveName(*ompEval).v;
else
return std::nullopt;
@@ -508,7 +487,7 @@ static void processHostEvalClauses(lower::AbstractConverter &converter,
HostEvalInfo *hostInfo = getHostEvalInfoStackTop(converter);
assert(hostInfo && "expected HOST_EVAL info structure");
- switch (extractOmpDirective(*ompEval)) {
+ switch (parser::omp::GetOmpDirectiveName(*ompEval).v) {
case OMPD_teams_distribute_parallel_do:
case OMPD_teams_distribute_parallel_do_simd:
cp.processThreadLimit(stmtCtx, hostInfo->ops);
@@ -569,7 +548,8 @@ static void processHostEvalClauses(lower::AbstractConverter &converter,
const auto *ompEval = eval.getIf<parser::OpenMPConstruct>();
assert(ompEval &&
- llvm::omp::allTargetSet.test(extractOmpDirective(*ompEval)) &&
+ llvm::omp::allTargetSet.test(
+ parser::omp::GetOmpDirectiveName(*ompEval).v) &&
"expected TARGET construct evaluation");
(void)ompEval;
@@ -3872,7 +3852,7 @@ static void genOMP(lower::AbstractConverter &converter, lower::SymMap &symTable,
std::get_if<common::Indirection<parser::OpenMPLoopConstruct>>(
&*optLoopCons)}) {
llvm::omp::Directive nestedDirective =
- getOpenMPDirectiveEnum(*ompNestedLoopCons);
+ parser::omp::GetOmpDirectiveName(*ompNestedLoopCons).v;
switch (nestedDirective) {
case llvm::omp::Directive::OMPD_tile:
// Emit the omp.loop_nest with annotation for tiling
@@ -3889,7 +3869,8 @@ static void genOMP(lower::AbstractConverter &converter, lower::SymMap &symTable,
}
}
- llvm::omp::Directive directive = getOpenMPDirectiveEnum(beginLoopDirective);
+ llvm::omp::Directive directive =
+ parser::omp::GetOmpDirectiveName(beginLoopDirective).v;
const parser::CharBlock &source =
std::get<parser::OmpLoopDirective>(beginLoopDirective.t).source;
ConstructQueue queue{
diff --git a/flang/lib/Lower/OpenMP/Utils.cpp b/flang/lib/Lower/OpenMP/Utils.cpp
index b1716d6..13fda97 100644
--- a/flang/lib/Lower/OpenMP/Utils.cpp
+++ b/flang/lib/Lower/OpenMP/Utils.cpp
@@ -20,6 +20,7 @@
#include <flang/Lower/PFTBuilder.h>
#include <flang/Optimizer/Builder/FIRBuilder.h>
#include <flang/Optimizer/Builder/Todo.h>
+#include <flang/Parser/openmp-utils.h>
#include <flang/Parser/parse-tree.h>
#include <flang/Parser/tools.h>
#include <flang/Semantics/tools.h>
@@ -663,89 +664,6 @@ bool collectLoopRelatedInfo(
return found;
}
-/// Get the directive enumeration value corresponding to the given OpenMP
-/// construct PFT node.
-llvm::omp::Directive
-extractOmpDirective(const parser::OpenMPConstruct &ompConstruct) {
- return common::visit(
- common::visitors{
- [](const parser::OpenMPAllocatorsConstruct &c) {
- return llvm::omp::OMPD_allocators;
- },
- [](const parser::OpenMPAssumeConstruct &c) {
- return llvm::omp::OMPD_assume;
- },
- [](const parser::OpenMPAtomicConstruct &c) {
- return llvm::omp::OMPD_atomic;
- },
- [](const parser::OpenMPBlockConstruct &c) {
- return std::get<parser::OmpBlockDirective>(
- std::get<parser::OmpBeginBlockDirective>(c.t).t)
- .v;
- },
- [](const parser::OpenMPCriticalConstruct &c) {
- return llvm::omp::OMPD_critical;
- },
- [](const parser::OpenMPDeclarativeAllocate &c) {
- return llvm::omp::OMPD_allocate;
- },
- [](const parser::OpenMPDispatchConstruct &c) {
- return llvm::omp::OMPD_dispatch;
- },
- [](const parser::OpenMPExecutableAllocate &c) {
- return llvm::omp::OMPD_allocate;
- },
- [](const parser::OpenMPLoopConstruct &c) {
- return std::get<parser::OmpLoopDirective>(
- std::get<parser::OmpBeginLoopDirective>(c.t).t)
- .v;
- },
- [](const parser::OpenMPSectionConstruct &c) {
- return llvm::omp::OMPD_section;
- },
- [](const parser::OpenMPSectionsConstruct &c) {
- return std::get<parser::OmpSectionsDirective>(
- std::get<parser::OmpBeginSectionsDirective>(c.t).t)
- .v;
- },
- [](const parser::OpenMPStandaloneConstruct &c) {
- return common::visit(
- common::visitors{
- [](const parser::OpenMPSimpleStandaloneConstruct &c) {
- return c.v.DirId();
- },
- [](const parser::OpenMPFlushConstruct &c) {
- return llvm::omp::OMPD_flush;
- },
- [](const parser::OpenMPCancelConstruct &c) {
- return llvm::omp::OMPD_cancel;
- },
- [](const parser::OpenMPCancellationPointConstruct &c) {
- return llvm::omp::OMPD_cancellation_point;
- },
- [](const parser::OmpMetadirectiveDirective &c) {
- return llvm::omp::OMPD_metadirective;
- },
- [](const parser::OpenMPDepobjConstruct &c) {
- return llvm::omp::OMPD_depobj;
- },
- [](const parser::OpenMPInteropConstruct &c) {
- return llvm::omp::OMPD_interop;
- }},
- c.u);
- },
- [](const parser::OpenMPUtilityConstruct &c) {
- return common::visit(
- common::visitors{[](const parser::OmpErrorDirective &c) {
- return llvm::omp::OMPD_error;
- },
- [](const parser::OmpNothingDirective &c) {
- return llvm::omp::OMPD_nothing;
- }},
- c.u);
- }},
- ompConstruct.u);
-}
} // namespace omp
} // namespace lower
} // namespace Fortran
diff --git a/flang/lib/Lower/OpenMP/Utils.h b/flang/lib/Lower/OpenMP/Utils.h
index 8e3ad5c..11641ba 100644
--- a/flang/lib/Lower/OpenMP/Utils.h
+++ b/flang/lib/Lower/OpenMP/Utils.h
@@ -167,8 +167,6 @@ bool collectLoopRelatedInfo(
mlir::omp::LoopRelatedClauseOps &result,
llvm::SmallVectorImpl<const semantics::Symbol *> &iv);
-llvm::omp::Directive
-extractOmpDirective(const parser::OpenMPConstruct &ompConstruct);
} // namespace omp
} // namespace lower
} // namespace Fortran
diff --git a/flang/lib/Lower/Support/ReductionProcessor.cpp b/flang/lib/Lower/Support/ReductionProcessor.cpp
index 80c32d0..605a5b6b 100644
--- a/flang/lib/Lower/Support/ReductionProcessor.cpp
+++ b/flang/lib/Lower/Support/ReductionProcessor.cpp
@@ -39,7 +39,7 @@ namespace lower {
namespace omp {
// explicit template declarations
-template void ReductionProcessor::processReductionArguments<
+template bool ReductionProcessor::processReductionArguments<
mlir::omp::DeclareReductionOp, omp::clause::ReductionOperatorList>(
mlir::Location currentLocation, lower::AbstractConverter &converter,
const omp::clause::ReductionOperatorList &redOperatorList,
@@ -48,7 +48,7 @@ template void ReductionProcessor::processReductionArguments<
llvm::SmallVectorImpl<mlir::Attribute> &reductionDeclSymbols,
const llvm::SmallVectorImpl<const semantics::Symbol *> &reductionSymbols);
-template void ReductionProcessor::processReductionArguments<
+template bool ReductionProcessor::processReductionArguments<
fir::DeclareReductionOp, llvm::SmallVector<fir::ReduceOperationEnum>>(
mlir::Location currentLocation, lower::AbstractConverter &converter,
const llvm::SmallVector<fir::ReduceOperationEnum> &redOperatorList,
@@ -607,7 +607,7 @@ static bool doReductionByRef(mlir::Value reductionVar) {
}
template <typename OpType, typename RedOperatorListTy>
-void ReductionProcessor::processReductionArguments(
+bool ReductionProcessor::processReductionArguments(
mlir::Location currentLocation, lower::AbstractConverter &converter,
const RedOperatorListTy &redOperatorList,
llvm::SmallVectorImpl<mlir::Value> &reductionVars,
@@ -627,10 +627,10 @@ void ReductionProcessor::processReductionArguments(
std::get_if<omp::clause::ProcedureDesignator>(&redOperator.u)) {
if (!ReductionProcessor::supportedIntrinsicProcReduction(
*reductionIntrinsic)) {
- return;
+ return false;
}
} else {
- return;
+ return false;
}
}
}
@@ -765,6 +765,8 @@ void ReductionProcessor::processReductionArguments(
if (isDoConcurrent)
builder.restoreInsertionPoint(dcIP);
+
+ return true;
}
const semantics::SourceName
diff --git a/flang/lib/Optimizer/Builder/IntrinsicCall.cpp b/flang/lib/Optimizer/Builder/IntrinsicCall.cpp
index 323d2fe..4753d0a 100644
--- a/flang/lib/Optimizer/Builder/IntrinsicCall.cpp
+++ b/flang/lib/Optimizer/Builder/IntrinsicCall.cpp
@@ -246,6 +246,7 @@ static constexpr IntrinsicHandler handlers[]{
{"abs", &I::genAbs},
{"achar", &I::genChar},
{"acosd", &I::genAcosd},
+ {"acospi", &I::genAcospi},
{"adjustl",
&I::genAdjustRtCall<fir::runtime::genAdjustL>,
{{{"string", asAddr}}},
@@ -278,6 +279,7 @@ static constexpr IntrinsicHandler handlers[]{
{{{"mask", asValue}, {"pred", asValue}}},
/*isElemental=*/false},
{"asind", &I::genAsind},
+ {"asinpi", &I::genAsinpi},
{"associated",
&I::genAssociated,
{{{"pointer", asInquired}, {"target", asInquired}}},
@@ -901,6 +903,7 @@ static constexpr IntrinsicHandler handlers[]{
{{{"number", asValue}, {"handler", asAddr}, {"status", asAddr}}},
/*isElemental=*/false},
{"sind", &I::genSind},
+ {"sinpi", &I::genSinpi},
{"size",
&I::genSize,
{{{"array", asBox},
@@ -941,6 +944,7 @@ static constexpr IntrinsicHandler handlers[]{
{{{"count", asAddr}, {"count_rate", asAddr}, {"count_max", asAddr}}},
/*isElemental=*/false},
{"tand", &I::genTand},
+ {"tanpi", &I::genTanpi},
{"this_grid", &I::genThisGrid, {}, /*isElemental=*/false},
{"this_thread_block", &I::genThisThreadBlock, {}, /*isElemental=*/false},
{"this_warp", &I::genThisWarp, {}, /*isElemental=*/false},
@@ -2674,6 +2678,21 @@ mlir::Value IntrinsicLibrary::genAcosd(mlir::Type resultType,
return mlir::arith::MulFOp::create(builder, loc, result, factor);
}
+// ACOSPI
+mlir::Value IntrinsicLibrary::genAcospi(mlir::Type resultType,
+ llvm::ArrayRef<mlir::Value> args) {
+ assert(args.size() == 1);
+ mlir::MLIRContext *context = builder.getContext();
+ mlir::FunctionType ftype =
+ mlir::FunctionType::get(context, {resultType}, {args[0].getType()});
+ mlir::Value acos = getRuntimeCallGenerator("acos", ftype)(builder, loc, args);
+ llvm::APFloat inv_pi = llvm::APFloat(llvm::numbers::inv_pi);
+ mlir::Value dfactor =
+ builder.createRealConstant(loc, mlir::Float64Type::get(context), inv_pi);
+ mlir::Value factor = builder.createConvert(loc, resultType, dfactor);
+ return mlir::arith::MulFOp::create(builder, loc, acos, factor);
+}
+
// ADJUSTL & ADJUSTR
template <void (*CallRuntime)(fir::FirOpBuilder &, mlir::Location loc,
mlir::Value, mlir::Value)>
@@ -2827,6 +2846,21 @@ mlir::Value IntrinsicLibrary::genAsind(mlir::Type resultType,
return mlir::arith::MulFOp::create(builder, loc, result, factor);
}
+// ASINPI
+mlir::Value IntrinsicLibrary::genAsinpi(mlir::Type resultType,
+ llvm::ArrayRef<mlir::Value> args) {
+ assert(args.size() == 1);
+ mlir::MLIRContext *context = builder.getContext();
+ mlir::FunctionType ftype =
+ mlir::FunctionType::get(context, {resultType}, {args[0].getType()});
+ mlir::Value asin = getRuntimeCallGenerator("asin", ftype)(builder, loc, args);
+ llvm::APFloat inv_pi = llvm::APFloat(llvm::numbers::inv_pi);
+ mlir::Value dfactor =
+ builder.createRealConstant(loc, mlir::Float64Type::get(context), inv_pi);
+ mlir::Value factor = builder.createConvert(loc, resultType, dfactor);
+ return mlir::arith::MulFOp::create(builder, loc, asin, factor);
+}
+
// ATAND, ATAN2D
mlir::Value IntrinsicLibrary::genAtand(mlir::Type resultType,
llvm::ArrayRef<mlir::Value> args) {
@@ -8079,6 +8113,21 @@ mlir::Value IntrinsicLibrary::genSind(mlir::Type resultType,
return getRuntimeCallGenerator("sin", ftype)(builder, loc, {arg});
}
+// SINPI
+mlir::Value IntrinsicLibrary::genSinpi(mlir::Type resultType,
+ llvm::ArrayRef<mlir::Value> args) {
+ assert(args.size() == 1);
+ mlir::MLIRContext *context = builder.getContext();
+ mlir::FunctionType ftype =
+ mlir::FunctionType::get(context, {resultType}, {args[0].getType()});
+ llvm::APFloat pi = llvm::APFloat(llvm::numbers::pi);
+ mlir::Value dfactor =
+ builder.createRealConstant(loc, mlir::Float64Type::get(context), pi);
+ mlir::Value factor = builder.createConvert(loc, args[0].getType(), dfactor);
+ mlir::Value arg = builder.create<mlir::arith::MulFOp>(loc, args[0], factor);
+ return getRuntimeCallGenerator("sin", ftype)(builder, loc, {arg});
+}
+
// SIZE
fir::ExtendedValue
IntrinsicLibrary::genSize(mlir::Type resultType,
@@ -8161,6 +8210,21 @@ mlir::Value IntrinsicLibrary::genTand(mlir::Type resultType,
return getRuntimeCallGenerator("tan", ftype)(builder, loc, {arg});
}
+// TANPI
+mlir::Value IntrinsicLibrary::genTanpi(mlir::Type resultType,
+ llvm::ArrayRef<mlir::Value> args) {
+ assert(args.size() == 1);
+ mlir::MLIRContext *context = builder.getContext();
+ mlir::FunctionType ftype =
+ mlir::FunctionType::get(context, {resultType}, {args[0].getType()});
+ llvm::APFloat pi = llvm::APFloat(llvm::numbers::pi);
+ mlir::Value dfactor =
+ builder.createRealConstant(loc, mlir::Float64Type::get(context), pi);
+ mlir::Value factor = builder.createConvert(loc, args[0].getType(), dfactor);
+ mlir::Value arg = builder.create<mlir::arith::MulFOp>(loc, args[0], factor);
+ return getRuntimeCallGenerator("tan", ftype)(builder, loc, {arg});
+}
+
// THIS_GRID
mlir::Value IntrinsicLibrary::genThisGrid(mlir::Type resultType,
llvm::ArrayRef<mlir::Value> args) {
diff --git a/flang/lib/Optimizer/Transforms/ExternalNameConversion.cpp b/flang/lib/Optimizer/Transforms/ExternalNameConversion.cpp
index 3d84eaa..2fcff87 100644
--- a/flang/lib/Optimizer/Transforms/ExternalNameConversion.cpp
+++ b/flang/lib/Optimizer/Transforms/ExternalNameConversion.cpp
@@ -41,6 +41,23 @@ mangleExternalName(const std::pair<fir::NameUniquer::NameKind,
appendUnderscore);
}
+/// Process a symbol reference and return the updated symbol reference if
+/// needed.
+std::optional<mlir::SymbolRefAttr>
+processSymbolRef(mlir::SymbolRefAttr symRef, mlir::Operation *nestedOp,
+ const llvm::DenseMap<mlir::StringAttr, mlir::FlatSymbolRefAttr>
+ &remappings) {
+ if (auto remap = remappings.find(symRef.getLeafReference());
+ remap != remappings.end()) {
+ mlir::SymbolRefAttr symAttr = mlir::FlatSymbolRefAttr(remap->second);
+ if (mlir::isa<mlir::gpu::LaunchFuncOp>(nestedOp))
+ symAttr = mlir::SymbolRefAttr::get(
+ symRef.getRootReference(), {mlir::FlatSymbolRefAttr(remap->second)});
+ return symAttr;
+ }
+ return std::nullopt;
+}
+
namespace {
class ExternalNameConversionPass
@@ -97,21 +114,40 @@ void ExternalNameConversionPass::runOnOperation() {
// Update all uses of the functions and globals that have been renamed.
op.walk([&remappings](mlir::Operation *nestedOp) {
- llvm::SmallVector<std::pair<mlir::StringAttr, mlir::SymbolRefAttr>> updates;
+ llvm::SmallVector<std::pair<mlir::StringAttr, mlir::SymbolRefAttr>>
+ symRefUpdates;
+ llvm::SmallVector<std::pair<mlir::StringAttr, mlir::ArrayAttr>>
+ arrayUpdates;
for (const mlir::NamedAttribute &attr : nestedOp->getAttrDictionary())
if (auto symRef = llvm::dyn_cast<mlir::SymbolRefAttr>(attr.getValue())) {
- if (auto remap = remappings.find(symRef.getLeafReference());
- remap != remappings.end()) {
- mlir::SymbolRefAttr symAttr = mlir::FlatSymbolRefAttr(remap->second);
- if (mlir::isa<mlir::gpu::LaunchFuncOp>(nestedOp))
- symAttr = mlir::SymbolRefAttr::get(
- symRef.getRootReference(),
- {mlir::FlatSymbolRefAttr(remap->second)});
- updates.emplace_back(std::pair<mlir::StringAttr, mlir::SymbolRefAttr>{
- attr.getName(), symAttr});
+ if (auto newSymRef = processSymbolRef(symRef, nestedOp, remappings))
+ symRefUpdates.emplace_back(
+ std::pair<mlir::StringAttr, mlir::SymbolRefAttr>{attr.getName(),
+ *newSymRef});
+ } else if (auto arrayAttr =
+ llvm::dyn_cast<mlir::ArrayAttr>(attr.getValue())) {
+ llvm::SmallVector<mlir::Attribute> symbolRefs;
+ for (auto element : arrayAttr) {
+ if (!element) {
+ symbolRefs.push_back(element);
+ continue;
+ }
+ auto symRef = llvm::dyn_cast<mlir::SymbolRefAttr>(element);
+ std::optional<mlir::SymbolRefAttr> updatedSymRef;
+ if (symRef)
+ updatedSymRef = processSymbolRef(symRef, nestedOp, remappings);
+ if (!symRef || !updatedSymRef)
+ symbolRefs.push_back(element);
+ else
+ symbolRefs.push_back(*updatedSymRef);
}
+ arrayUpdates.push_back(std::make_pair(
+ attr.getName(),
+ mlir::ArrayAttr::get(nestedOp->getContext(), symbolRefs)));
}
- for (auto update : updates)
+ for (auto update : symRefUpdates)
+ nestedOp->setAttr(update.first, update.second);
+ for (auto update : arrayUpdates)
nestedOp->setAttr(update.first, update.second);
});
}
diff --git a/flang/lib/Semantics/canonicalize-omp.cpp b/flang/lib/Semantics/canonicalize-omp.cpp
index 77e2fd6..9722eca 100644
--- a/flang/lib/Semantics/canonicalize-omp.cpp
+++ b/flang/lib/Semantics/canonicalize-omp.cpp
@@ -9,6 +9,7 @@
#include "canonicalize-omp.h"
#include "flang/Parser/parse-tree-visitor.h"
#include "flang/Parser/parse-tree.h"
+#include "flang/Semantics/semantics.h"
// After Loop Canonicalization, rewrite OpenMP parse tree to make OpenMP
// Constructs more structured which provide explicit scopes for later
generated by cgit v1.1 at 2025-07-24 08:05:04 +0000