[flang-commits] [flang] [flang][NFC] turn fir.call is_bind_c into enum for procedure flags (PR #105691)
via flang-commits
flang-commits at lists.llvm.org
Fri Aug 23 00:23:34 PDT 2024
https://github.com/jeanPerier updated https://github.com/llvm/llvm-project/pull/105691
>From 46358fe84b3d8cd1ef18968e55f8c139cbfb3955 Mon Sep 17 00:00:00 2001
From: Jean Perier <jperier at nvidia.com>
Date: Thu, 22 Aug 2024 08:25:20 -0700
Subject: [PATCH 1/2] [flang][NFC] turn fir.call is_bind_c into enum attrs with
procedure flags
First patch to fix a BIND(C) ABI issue
(https://github.com/llvm/llvm-project/issues/102113). I need to keep track of
BIND(C) in more locations: fir.dispatch and func.func, and I need to fix a few
passes that are dropping the attribute on the floor. Since I expect more
procedure attributes that cannot be reflected in mlir::FunctionType will be
needed for ABI, optimizations, or debug info, this NFC patch adds a new enum
attribute to keep track of procedure attributes in the IR.
This patch is not updating lowering to lower more attributes, this will be done
in a separate patch to keep the test change low here.
---
.../flang/Optimizer/Dialect/FIRAttr.td | 27 +++++++++++++++++++
.../include/flang/Optimizer/Dialect/FIROps.td | 4 +--
flang/lib/Lower/ConvertCall.cpp | 18 ++++++++-----
flang/lib/Lower/ConvertExpr.cpp | 4 +--
flang/lib/Optimizer/Builder/IntrinsicCall.cpp | 4 +--
flang/lib/Optimizer/Dialect/FIRAttr.cpp | 3 ++-
flang/lib/Optimizer/Dialect/FIROps.cpp | 24 ++++++++++++++---
.../ConstantArgumentGlobalisation.cpp | 4 +--
.../Transforms/PolymorphicOpConversion.cpp | 5 +++-
.../lib/Optimizer/Transforms/StackArrays.cpp | 8 +++---
flang/test/HLFIR/c_ptr_byvalue.f90 | 4 +--
flang/test/Lower/CUDA/cuda-device-proc.cuf | 8 +++---
flang/test/Lower/HLFIR/assumed-rank-calls.f90 | 2 +-
flang/test/Lower/HLFIR/assumed-rank-iface.f90 | 4 +--
.../test/Lower/HLFIR/bindc-value-derived.f90 | 2 +-
flang/test/Lower/HLFIR/block_bindc_pocs.f90 | 2 +-
.../call-sequence-associated-descriptors.f90 | 8 +++---
.../calls-character-singleton-result.f90 | 2 +-
.../Lower/HLFIR/ignore-type-assumed-shape.f90 | 8 +++---
.../OpenMP/threadprivate-default-clause.f90 | 4 +--
flang/test/Lower/block.f90 | 2 +-
flang/test/Lower/call-bindc.f90 | 2 +-
22 files changed, 101 insertions(+), 48 deletions(-)
diff --git a/flang/include/flang/Optimizer/Dialect/FIRAttr.td b/flang/include/flang/Optimizer/Dialect/FIRAttr.td
index 60281dfa637139..6400756b384482 100644
--- a/flang/include/flang/Optimizer/Dialect/FIRAttr.td
+++ b/flang/include/flang/Optimizer/Dialect/FIRAttr.td
@@ -58,6 +58,33 @@ def fir_FortranVariableFlagsAttr : fir_Attr<"FortranVariableFlags"> {
"::fir::FortranVariableFlagsAttr::get($_builder.getContext(), $0)";
}
+
+/// Fortran procedure attributes (F2023 15.6.2.1). BIND attribute (18.3.7)
+/// is also tracked in the same enum. Recursive (resp. Impure) attribute
+/// is implied by the absence of opposite NonRecursive (resp. Pure) attribute.
+def FIRfuncNoAttributes : I32BitEnumAttrCaseNone<"none">;
+def FIRfuncElemental : I32BitEnumAttrCaseBit<"elemental", 0>;
+def FIRfuncPure : I32BitEnumAttrCaseBit<"pure", 1>;
+def FIRfuncNonRecursive : I32BitEnumAttrCaseBit<"non_recursive", 2>;
+def FIRfuncSimple : I32BitEnumAttrCaseBit<"simple", 3>;
+def FIRfuncBind_c : I32BitEnumAttrCaseBit<"bind_c", 4>;
+
+def fir_FortranProcedureFlagsEnum : I32BitEnumAttr<
+ "FortranProcedureFlagsEnum",
+ "Fortran procedure attributes",
+ [FIRfuncNoAttributes, FIRfuncElemental, FIRfuncPure, FIRfuncNonRecursive,
+ FIRfuncSimple, FIRfuncBind_c]> {
+ let separator = ", ";
+ let cppNamespace = "::fir";
+ let genSpecializedAttr = 0;
+ let printBitEnumPrimaryGroups = 1;
+}
+
+def fir_FortranProcedureFlagsAttr :
+ EnumAttr<FIROpsDialect, fir_FortranProcedureFlagsEnum, "proc_attrs"> {
+ let assemblyFormat = "`<` $value `>`";
+}
+
def fir_BoxFieldAttr : I32EnumAttr<
"BoxFieldAttr", "",
[
diff --git a/flang/include/flang/Optimizer/Dialect/FIROps.td b/flang/include/flang/Optimizer/Dialect/FIROps.td
index 04f5fedf2783fc..2cd202c3623250 100644
--- a/flang/include/flang/Optimizer/Dialect/FIROps.td
+++ b/flang/include/flang/Optimizer/Dialect/FIROps.td
@@ -2432,9 +2432,9 @@ def fir_CallOp : fir_Op<"call",
let arguments = (ins
OptionalAttr<SymbolRefAttr>:$callee,
Variadic<AnyType>:$args,
+ OptionalAttr<fir_FortranProcedureFlagsAttr>:$procedure_attrs,
DefaultValuedAttr<Arith_FastMathAttr,
- "::mlir::arith::FastMathFlags::none">:$fastmath,
- UnitAttr:$is_bind_c
+ "::mlir::arith::FastMathFlags::none">:$fastmath
);
let results = (outs Variadic<AnyType>);
diff --git a/flang/lib/Lower/ConvertCall.cpp b/flang/lib/Lower/ConvertCall.cpp
index fd873f55dd844e..f445a21e560bc9 100644
--- a/flang/lib/Lower/ConvertCall.cpp
+++ b/flang/lib/Lower/ConvertCall.cpp
@@ -372,7 +372,7 @@ std::pair<fir::ExtendedValue, bool> Fortran::lower::genCallOpAndResult(
auto stackSaveSymbol = bldr->getSymbolRefAttr(stackSaveFn.getName());
mlir::Value sp;
fir::CallOp call = bldr->create<fir::CallOp>(
- loc, stackSaveFn.getFunctionType().getResults(), stackSaveSymbol,
+ loc, stackSaveSymbol, stackSaveFn.getFunctionType().getResults(),
mlir::ValueRange{});
if (call.getNumResults() != 0)
sp = call.getResult(0);
@@ -380,9 +380,9 @@ std::pair<fir::ExtendedValue, bool> Fortran::lower::genCallOpAndResult(
auto stackRestoreFn = fir::factory::getLlvmStackRestore(*bldr);
auto stackRestoreSymbol =
bldr->getSymbolRefAttr(stackRestoreFn.getName());
- bldr->create<fir::CallOp>(loc,
+ bldr->create<fir::CallOp>(loc, stackRestoreSymbol,
stackRestoreFn.getFunctionType().getResults(),
- stackRestoreSymbol, mlir::ValueRange{sp});
+ mlir::ValueRange{sp});
});
}
mlir::Value temp =
@@ -640,11 +640,15 @@ std::pair<fir::ExtendedValue, bool> Fortran::lower::genCallOpAndResult(
if (callNumResults != 0)
callResult = dispatch.getResult(0);
} else {
- // Standard procedure call with fir.call.
- auto call = builder.create<fir::CallOp>(loc, funcType.getResults(),
- funcSymbolAttr, operands);
+ // TODO: gather other procedure attributes.
+ fir::FortranProcedureFlagsEnumAttr procAttrs;
if (caller.characterize().IsBindC())
- call.setIsBindC(true);
+ procAttrs = fir::FortranProcedureFlagsEnumAttr::get(
+ builder.getContext(), fir::FortranProcedureFlagsEnum::bind_c);
+
+ // Standard procedure call with fir.call.
+ auto call = builder.create<fir::CallOp>(
+ loc, funcType.getResults(), funcSymbolAttr, operands, procAttrs);
callNumResults = call.getNumResults();
if (callNumResults != 0)
diff --git a/flang/lib/Lower/ConvertExpr.cpp b/flang/lib/Lower/ConvertExpr.cpp
index 44c3dc88edd32f..7dd317d64436b5 100644
--- a/flang/lib/Lower/ConvertExpr.cpp
+++ b/flang/lib/Lower/ConvertExpr.cpp
@@ -6120,7 +6120,7 @@ class ArrayExprLowering {
mlir::SymbolRefAttr funcSymAttr =
builder.getSymbolRefAttr(memcpyFunc.getName());
mlir::FunctionType funcTy = memcpyFunc.getFunctionType();
- builder.create<fir::CallOp>(loc, funcTy.getResults(), funcSymAttr, args);
+ builder.create<fir::CallOp>(loc, funcSymAttr, funcTy.getResults(), args);
}
// Construct code to check for a buffer overrun and realloc the buffer when
@@ -6146,7 +6146,7 @@ class ArrayExprLowering {
builder.getSymbolRefAttr(reallocFunc.getName());
mlir::FunctionType funcTy = reallocFunc.getFunctionType();
auto newMem = builder.create<fir::CallOp>(
- loc, funcTy.getResults(), funcSymAttr,
+ loc, funcSymAttr, funcTy.getResults(),
llvm::ArrayRef<mlir::Value>{
builder.createConvert(loc, funcTy.getInputs()[0], mem),
builder.createConvert(loc, funcTy.getInputs()[1], byteSz)});
diff --git a/flang/lib/Optimizer/Builder/IntrinsicCall.cpp b/flang/lib/Optimizer/Builder/IntrinsicCall.cpp
index 22439010e7797b..dc0dc47bda9a9d 100644
--- a/flang/lib/Optimizer/Builder/IntrinsicCall.cpp
+++ b/flang/lib/Optimizer/Builder/IntrinsicCall.cpp
@@ -819,8 +819,8 @@ mlir::Value genLibCall(fir::FirOpBuilder &builder, mlir::Location loc,
llvm::SmallVector<mlir::Value, 3> operands{funcPointer};
operands.append(args.begin(), args.end());
- libCall = builder.create<fir::CallOp>(loc, libFuncType.getResults(),
- nullptr, operands);
+ libCall = builder.create<fir::CallOp>(loc, mlir::SymbolRefAttr{},
+ libFuncType.getResults(), operands);
}
LLVM_DEBUG(libCall.dump(); llvm::dbgs() << "\n");
diff --git a/flang/lib/Optimizer/Dialect/FIRAttr.cpp b/flang/lib/Optimizer/Dialect/FIRAttr.cpp
index 443e94ae6606f1..4c78e223b41785 100644
--- a/flang/lib/Optimizer/Dialect/FIRAttr.cpp
+++ b/flang/lib/Optimizer/Dialect/FIRAttr.cpp
@@ -296,7 +296,8 @@ void fir::printFirAttribute(FIROpsDialect *dialect, mlir::Attribute attr,
//===----------------------------------------------------------------------===//
void FIROpsDialect::registerAttributes() {
- addAttributes<ClosedIntervalAttr, ExactTypeAttr, FortranVariableFlagsAttr,
+ addAttributes<ClosedIntervalAttr, ExactTypeAttr,
+ FortranProcedureFlagsEnumAttr, FortranVariableFlagsAttr,
LowerBoundAttr, PointIntervalAttr, RealAttr, ReduceAttr,
SubclassAttr, UpperBoundAttr, LocationKindAttr,
LocationKindArrayAttr>();
diff --git a/flang/lib/Optimizer/Dialect/FIROps.cpp b/flang/lib/Optimizer/Dialect/FIROps.cpp
index 9e6b88041ba69d..c2a893aee7b0a5 100644
--- a/flang/lib/Optimizer/Dialect/FIROps.cpp
+++ b/flang/lib/Optimizer/Dialect/FIROps.cpp
@@ -1103,6 +1103,15 @@ void fir::CallOp::print(mlir::OpAsmPrinter &p) {
p << getOperand(0);
p << '(' << (*this)->getOperands().drop_front(isDirect ? 0 : 1) << ')';
+ // Print `proc_attrs<...>`, if present.
+ fir::FortranProcedureFlagsEnumAttr procAttrs = getProcedureAttrsAttr();
+ if (procAttrs &&
+ procAttrs.getValue() != fir::FortranProcedureFlagsEnum::none) {
+ p << ' ' << fir::FortranProcedureFlagsEnumAttr::getMnemonic();
+ mlir::arith::FastMathFlagsAttr::getMnemonic();
+ p.printStrippedAttrOrType(procAttrs);
+ }
+
// Print 'fastmath<...>' (if it has non-default value) before
// any other attributes.
mlir::arith::FastMathFlagsAttr fmfAttr = getFastmathAttr();
@@ -1111,9 +1120,9 @@ void fir::CallOp::print(mlir::OpAsmPrinter &p) {
p.printStrippedAttrOrType(fmfAttr);
}
- p.printOptionalAttrDict(
- (*this)->getAttrs(),
- {fir::CallOp::getCalleeAttrNameStr(), getFastmathAttrName()});
+ p.printOptionalAttrDict((*this)->getAttrs(),
+ {fir::CallOp::getCalleeAttrNameStr(),
+ getFastmathAttrName(), getProcedureAttrsAttrName()});
auto resultTypes{getResultTypes()};
llvm::SmallVector<mlir::Type> argTypes(
llvm::drop_begin(getOperandTypes(), isDirect ? 0 : 1));
@@ -1138,6 +1147,15 @@ mlir::ParseResult fir::CallOp::parse(mlir::OpAsmParser &parser,
if (parser.parseOperandList(operands, mlir::OpAsmParser::Delimiter::Paren))
return mlir::failure();
+ // Parse `proc_attrs<...>`, if present.
+ fir::FortranProcedureFlagsEnumAttr procAttr;
+ if (mlir::succeeded(parser.parseOptionalKeyword(
+ fir::FortranProcedureFlagsEnumAttr::getMnemonic())))
+ if (parser.parseCustomAttributeWithFallback(
+ procAttr, mlir::Type{}, getProcedureAttrsAttrName(result.name),
+ attrs))
+ return mlir::failure();
+
// Parse 'fastmath<...>', if present.
mlir::arith::FastMathFlagsAttr fmfAttr;
llvm::StringRef fmfAttrName = getFastmathAttrName(result.name);
diff --git a/flang/lib/Optimizer/Transforms/ConstantArgumentGlobalisation.cpp b/flang/lib/Optimizer/Transforms/ConstantArgumentGlobalisation.cpp
index 1e44288c784c2a..eef6f047fc1bf8 100644
--- a/flang/lib/Optimizer/Transforms/ConstantArgumentGlobalisation.cpp
+++ b/flang/lib/Optimizer/Transforms/ConstantArgumentGlobalisation.cpp
@@ -126,10 +126,10 @@ class CallOpRewriter : public mlir::OpRewritePattern<fir::CallOp> {
newResultTypes.append(callOp.getResultTypes().begin(),
callOp.getResultTypes().end());
fir::CallOp newOp = builder.create<fir::CallOp>(
- loc, newResultTypes,
+ loc,
callOp.getCallee().has_value() ? callOp.getCallee().value()
: mlir::SymbolRefAttr{},
- newOperands);
+ newResultTypes, newOperands);
// Copy all the attributes from the old to new op.
newOp->setAttrs(callOp->getAttrs());
rewriter.replaceOp(callOp, newOp);
diff --git a/flang/lib/Optimizer/Transforms/PolymorphicOpConversion.cpp b/flang/lib/Optimizer/Transforms/PolymorphicOpConversion.cpp
index 57f19f257b569d..105f275de8b940 100644
--- a/flang/lib/Optimizer/Transforms/PolymorphicOpConversion.cpp
+++ b/flang/lib/Optimizer/Transforms/PolymorphicOpConversion.cpp
@@ -205,7 +205,10 @@ struct DispatchOpConv : public OpConversionPattern<fir::DispatchOp> {
// Make the call.
llvm::SmallVector<mlir::Value> args{funcPtr};
args.append(dispatch.getArgs().begin(), dispatch.getArgs().end());
- rewriter.replaceOpWithNewOp<fir::CallOp>(dispatch, resTypes, nullptr, args);
+ // FIXME: add procedure_attrs to fir.dispatch and propagate to fir.call.
+ rewriter.replaceOpWithNewOp<fir::CallOp>(
+ dispatch, resTypes, nullptr, args,
+ /*procedure_attrs=*/fir::FortranProcedureFlagsEnumAttr{});
return mlir::success();
}
diff --git a/flang/lib/Optimizer/Transforms/StackArrays.cpp b/flang/lib/Optimizer/Transforms/StackArrays.cpp
index a5a95138ac1281..e2b9b08a10ee76 100644
--- a/flang/lib/Optimizer/Transforms/StackArrays.cpp
+++ b/flang/lib/Optimizer/Transforms/StackArrays.cpp
@@ -738,9 +738,9 @@ void AllocMemConversion::insertStackSaveRestore(
builder.setInsertionPoint(oldAlloc);
mlir::Value sp =
builder
- .create<fir::CallOp>(oldAlloc.getLoc(),
+ .create<fir::CallOp>(oldAlloc.getLoc(), stackSaveSym,
stackSaveFn.getFunctionType().getResults(),
- stackSaveSym, mlir::ValueRange{})
+ mlir::ValueRange{})
.getResult(0);
mlir::func::FuncOp stackRestoreFn =
@@ -750,9 +750,9 @@ void AllocMemConversion::insertStackSaveRestore(
auto createStackRestoreCall = [&](mlir::Operation *user) {
builder.setInsertionPoint(user);
- builder.create<fir::CallOp>(user->getLoc(),
+ builder.create<fir::CallOp>(user->getLoc(), stackRestoreSym,
stackRestoreFn.getFunctionType().getResults(),
- stackRestoreSym, mlir::ValueRange{sp});
+ mlir::ValueRange{sp});
};
for (mlir::Operation *user : oldAlloc->getUsers()) {
diff --git a/flang/test/HLFIR/c_ptr_byvalue.f90 b/flang/test/HLFIR/c_ptr_byvalue.f90
index ea48bdb1a100f0..b2c8da5e22579d 100644
--- a/flang/test/HLFIR/c_ptr_byvalue.f90
+++ b/flang/test/HLFIR/c_ptr_byvalue.f90
@@ -7,7 +7,7 @@
! CHECK: %[[VAL_113:.*]] = fir.load %[[VAL_112]] : !fir.ref<i64>
! CHECK: %[[VAL_114:.*]] = fir.convert %[[VAL_113]] : (i64) -> !fir.ref<i64>
! CHECK: hlfir.end_associate %[[VAL_110]]#1, %[[VAL_110]]#2 : !fir.ref<!fir.type<_QM__fortran_builtinsT__builtin_c_ptr{__address:i64}>>, i1
-! CHECK: fir.call @get_expected_f(%[[VAL_114]]) fastmath<contract> {is_bind_c} : (!fir.ref<i64>) -> ()
+! CHECK: fir.call @get_expected_f(%[[VAL_114]]) proc_attrs<bind_c> fastmath<contract> : (!fir.ref<i64>) -> ()
subroutine test1
use iso_c_binding
interface
@@ -28,7 +28,7 @@ end subroutine get_expected_f
! CHECK: %[[VAL_99:.*]] = fir.coordinate_of %[[VAL_97]]#0, %[[VAL_98]] : (!fir.ref<!fir.type<_QM__fortran_builtinsT__builtin_c_ptr{__address:i64}>>, !fir.field) -> !fir.ref<i64>
! CHECK: %[[VAL_100:.*]] = fir.load %[[VAL_99]] : !fir.ref<i64>
! CHECK: %[[VAL_101:.*]] = fir.convert %[[VAL_100]] : (i64) -> !fir.ref<i64>
-! CHECK: fir.call @get_expected_f(%[[VAL_101]]) fastmath<contract> {is_bind_c} : (!fir.ref<i64>) -> ()
+! CHECK: fir.call @get_expected_f(%[[VAL_101]]) proc_attrs<bind_c> fastmath<contract> : (!fir.ref<i64>) -> ()
subroutine test2(cptr)
use iso_c_binding
interface
diff --git a/flang/test/Lower/CUDA/cuda-device-proc.cuf b/flang/test/Lower/CUDA/cuda-device-proc.cuf
index e259e6ee00f91e..bed0a4574fe94d 100644
--- a/flang/test/Lower/CUDA/cuda-device-proc.cuf
+++ b/flang/test/Lower/CUDA/cuda-device-proc.cuf
@@ -18,13 +18,13 @@ end
! CHECK-LABEL: func.func @_QPdevsub() attributes {cuf.proc_attr = #cuf.cuda_proc<global>}
! CHECK: fir.call @__syncthreads()
-! CHECK: fir.call @__syncwarp(%{{.*}}) fastmath<contract> {is_bind_c} : (!fir.ref<i32>) -> ()
+! CHECK: fir.call @__syncwarp(%{{.*}}) proc_attrs<bind_c> fastmath<contract> : (!fir.ref<i32>) -> ()
! CHECK: fir.call @__threadfence()
! CHECK: fir.call @__threadfence_block()
! CHECK: fir.call @__threadfence_system()
-! CHECK: %{{.*}} = fir.call @__syncthreads_and(%{{.*}}) fastmath<contract> {is_bind_c} : (!fir.ref<i32>) -> i32
-! CHECK: %{{.*}} = fir.call @__syncthreads_count(%{{.*}}) fastmath<contract> {is_bind_c} : (!fir.ref<i32>) -> i32
-! CHECK: %{{.*}} = fir.call @__syncthreads_or(%{{.*}}) fastmath<contract> {is_bind_c} : (!fir.ref<i32>) -> i32
+! CHECK: %{{.*}} = fir.call @__syncthreads_and(%{{.*}}) proc_attrs<bind_c> fastmath<contract> : (!fir.ref<i32>) -> i32
+! CHECK: %{{.*}} = fir.call @__syncthreads_count(%{{.*}}) proc_attrs<bind_c> fastmath<contract> : (!fir.ref<i32>) -> i32
+! CHECK: %{{.*}} = fir.call @__syncthreads_or(%{{.*}}) proc_attrs<bind_c> fastmath<contract> : (!fir.ref<i32>) -> i32
! CHECK: func.func private @__syncthreads() attributes {cuf.proc_attr = #cuf.cuda_proc<device>, fir.bindc_name = "__syncthreads"}
! CHECK: func.func private @__syncwarp(!fir.ref<i32> {cuf.data_attr = #cuf.cuda<device>}) attributes {cuf.proc_attr = #cuf.cuda_proc<device>, fir.bindc_name = "__syncwarp"}
diff --git a/flang/test/Lower/HLFIR/assumed-rank-calls.f90 b/flang/test/Lower/HLFIR/assumed-rank-calls.f90
index 9d4503fef6fce9..afb2bbac998910 100644
--- a/flang/test/Lower/HLFIR/assumed-rank-calls.f90
+++ b/flang/test/Lower/HLFIR/assumed-rank-calls.f90
@@ -36,7 +36,7 @@ subroutine bindc_func(x) bind(c)
! CHECK: %[[VAL_1:.*]] = fir.dummy_scope : !fir.dscope
! CHECK: %[[VAL_2:.*]]:2 = hlfir.declare %[[VAL_0]] dummy_scope %[[VAL_1]] {uniq_name = "_QFtest_to_bindcEx"} : (!fir.box<!fir.array<*:f32>>, !fir.dscope) -> (!fir.box<!fir.array<*:f32>>, !fir.box<!fir.array<*:f32>>)
! CHECK: %[[VAL_3:.*]] = fir.rebox_assumed_rank %[[VAL_2]]#0 lbs zeroes : (!fir.box<!fir.array<*:f32>>) -> !fir.box<!fir.array<*:f32>>
-! CHECK: fir.call @bindc_func(%[[VAL_3]]) fastmath<contract> {is_bind_c} : (!fir.box<!fir.array<*:f32>>) -> ()
+! CHECK: fir.call @bindc_func(%[[VAL_3]]) proc_attrs<bind_c> fastmath<contract> : (!fir.box<!fir.array<*:f32>>) -> ()
! CHECK: return
! CHECK: }
diff --git a/flang/test/Lower/HLFIR/assumed-rank-iface.f90 b/flang/test/Lower/HLFIR/assumed-rank-iface.f90
index aaf003a59d84f1..0e094cc6646d1f 100644
--- a/flang/test/Lower/HLFIR/assumed-rank-iface.f90
+++ b/flang/test/Lower/HLFIR/assumed-rank-iface.f90
@@ -38,7 +38,7 @@ subroutine int_scalar_to_assumed_rank_bindc(x)
! CHECK: %[[VAL_1:.*]]:2 = hlfir.declare %[[VAL_0]] dummy_scope %{{[0-9]+}} {uniq_name = "_QFint_scalar_to_assumed_rank_bindcEx"} : (!fir.ref<i32>, !fir.dscope) -> (!fir.ref<i32>, !fir.ref<i32>)
! CHECK: %[[VAL_2:.*]] = fir.embox %[[VAL_1]]#0 : (!fir.ref<i32>) -> !fir.box<i32>
! CHECK: %[[VAL_3:.*]] = fir.convert %[[VAL_2]] : (!fir.box<i32>) -> !fir.box<!fir.array<*:i32>>
-! CHECK: fir.call @int_assumed_rank_bindc(%[[VAL_3]]) fastmath<contract> {is_bind_c} : (!fir.box<!fir.array<*:i32>>) -> ()
+! CHECK: fir.call @int_assumed_rank_bindc(%[[VAL_3]]) proc_attrs<bind_c> fastmath<contract> : (!fir.box<!fir.array<*:i32>>) -> ()
subroutine int_r1_to_assumed_rank(x)
use ifaces, only : int_assumed_rank
@@ -94,7 +94,7 @@ subroutine int_assumed_shape_to_assumed_rank_bindc(x)
! CHECK: %[[VAL_3:.*]] = fir.shift %[[VAL_2]], %[[VAL_2]] : (index, index) -> !fir.shift<2>
! CHECK: %[[VAL_4:.*]] = fir.rebox %[[VAL_1]]#0(%[[VAL_3]]) : (!fir.box<!fir.array<?x?xi32>>, !fir.shift<2>) -> !fir.box<!fir.array<?x?xi32>>
! CHECK: %[[VAL_5:.*]] = fir.convert %[[VAL_4]] : (!fir.box<!fir.array<?x?xi32>>) -> !fir.box<!fir.array<*:i32>>
-! CHECK: fir.call @int_assumed_rank_bindc(%[[VAL_5]]) fastmath<contract> {is_bind_c} : (!fir.box<!fir.array<*:i32>>) -> ()
+! CHECK: fir.call @int_assumed_rank_bindc(%[[VAL_5]]) proc_attrs<bind_c> fastmath<contract> : (!fir.box<!fir.array<*:i32>>) -> ()
subroutine int_allocatable_to_assumed_rank(x)
use ifaces, only : int_assumed_rank
diff --git a/flang/test/Lower/HLFIR/bindc-value-derived.f90 b/flang/test/Lower/HLFIR/bindc-value-derived.f90
index 8cefd1246173c2..a54b29b470e0b4 100644
--- a/flang/test/Lower/HLFIR/bindc-value-derived.f90
+++ b/flang/test/Lower/HLFIR/bindc-value-derived.f90
@@ -31,7 +31,7 @@ subroutine call_it(x)
! CHECK-SAME: %[[VAL_0:.*]]: !fir.ref<!fir.type<_QMbindc_byvalTt{i:i32}>> {fir.bindc_name = "x"}) {
! CHECK: %[[VAL_1:.*]]:2 = hlfir.declare %[[VAL_0]] dummy_scope %{{[0-9]+}} {uniq_name = "_QMbindc_byvalFcall_itEx"} : (!fir.ref<!fir.type<_QMbindc_byvalTt{i:i32}>>, !fir.dscope) -> (!fir.ref<!fir.type<_QMbindc_byvalTt{i:i32}>>, !fir.ref<!fir.type<_QMbindc_byvalTt{i:i32}>>)
! CHECK: %[[VAL_2:.*]] = fir.load %[[VAL_1]]#1 : !fir.ref<!fir.type<_QMbindc_byvalTt{i:i32}>>
-! CHECK: fir.call @test(%[[VAL_2]]) fastmath<contract> {is_bind_c} : (!fir.type<_QMbindc_byvalTt{i:i32}>) -> ()
+! CHECK: fir.call @test(%[[VAL_2]]) proc_attrs<bind_c> fastmath<contract> : (!fir.type<_QMbindc_byvalTt{i:i32}>) -> ()
! CHECK: return
! CHECK: }
end module
diff --git a/flang/test/Lower/HLFIR/block_bindc_pocs.f90 b/flang/test/Lower/HLFIR/block_bindc_pocs.f90
index c6519e92d0b5cd..090eeb35ea88b0 100644
--- a/flang/test/Lower/HLFIR/block_bindc_pocs.f90
+++ b/flang/test/Lower/HLFIR/block_bindc_pocs.f90
@@ -9,7 +9,7 @@ end subroutine test_proc
end interface
end module m
!CHECK-DAG: %[[S0:.*]] = fir.call @llvm.stacksave.p0() fastmath<contract> : () -> !fir.ref<i8>
-!CHECK-DAG: fir.call @test_proc() fastmath<contract> {is_bind_c} : () -> ()
+!CHECK-DAG: fir.call @test_proc() proc_attrs<bind_c> fastmath<contract> : () -> ()
!CHECK-DAG: fir.call @llvm.stackrestore.p0(%[[S0]]) fastmath<contract> : (!fir.ref<i8>) -> ()
!CHECK-DAG: func.func private @test_proc() attributes {fir.bindc_name = "test_proc"}
subroutine test
diff --git a/flang/test/Lower/HLFIR/call-sequence-associated-descriptors.f90 b/flang/test/Lower/HLFIR/call-sequence-associated-descriptors.f90
index ccbc1df96a73a5..86464f29e0831b 100644
--- a/flang/test/Lower/HLFIR/call-sequence-associated-descriptors.f90
+++ b/flang/test/Lower/HLFIR/call-sequence-associated-descriptors.f90
@@ -46,7 +46,7 @@ subroutine test_char_1(x)
! CHECK: %[[VAL_26:.*]] = fir.box_addr %[[VAL_10]] : (!fir.box<!fir.array<10x20x!fir.char<1,?>>>) -> !fir.ref<!fir.array<10x20x!fir.char<1,?>>>
! CHECK: %[[VAL_27:.*]] = fir.convert %[[VAL_26]] : (!fir.ref<!fir.array<10x20x!fir.char<1,?>>>) -> !fir.ref<!fir.array<?x!fir.char<1,?>>>
! CHECK: %[[VAL_28:.*]] = fir.embox %[[VAL_27]](%[[VAL_24]]) typeparams %[[VAL_25]] : (!fir.ref<!fir.array<?x!fir.char<1,?>>>, !fir.shapeshift<1>, index) -> !fir.box<!fir.array<?x!fir.char<1,?>>>
-! CHECK: fir.call @takes_char(%[[VAL_28]], %[[VAL_11]]#1) fastmath<contract> {is_bind_c} : (!fir.box<!fir.array<?x!fir.char<1,?>>>, !fir.ref<i32>) -> ()
+! CHECK: fir.call @takes_char(%[[VAL_28]], %[[VAL_11]]#1) proc_attrs<bind_c> fastmath<contract> : (!fir.box<!fir.array<?x!fir.char<1,?>>>, !fir.ref<i32>) -> ()
! CHECK: hlfir.end_associate %[[VAL_11]]#1, %[[VAL_11]]#2 : !fir.ref<i32>, i1
! CHECK: return
! CHECK: }
@@ -80,7 +80,7 @@ subroutine test_char_copy_in_copy_out(x)
! CHECK: %[[VAL_22:.*]] = fir.box_addr %[[VAL_6]] : (!fir.box<!fir.array<?x?x!fir.char<1,?>>>) -> !fir.ref<!fir.array<?x?x!fir.char<1,?>>>
! CHECK: %[[VAL_23:.*]] = fir.convert %[[VAL_22]] : (!fir.ref<!fir.array<?x?x!fir.char<1,?>>>) -> !fir.ref<!fir.array<?x!fir.char<1,?>>>
! CHECK: %[[VAL_24:.*]] = fir.embox %[[VAL_23]](%[[VAL_20]]) typeparams %[[VAL_21]] : (!fir.ref<!fir.array<?x!fir.char<1,?>>>, !fir.shapeshift<1>, index) -> !fir.box<!fir.array<?x!fir.char<1,?>>>
-! CHECK: fir.call @takes_char(%[[VAL_24]], %[[VAL_7]]#1) fastmath<contract> {is_bind_c} : (!fir.box<!fir.array<?x!fir.char<1,?>>>, !fir.ref<i32>) -> ()
+! CHECK: fir.call @takes_char(%[[VAL_24]], %[[VAL_7]]#1) proc_attrs<bind_c> fastmath<contract> : (!fir.box<!fir.array<?x!fir.char<1,?>>>, !fir.ref<i32>) -> ()
! CHECK: hlfir.copy_out %[[TMP_BOX]], %[[VAL_3]]#1 to %[[VAL_1]]#0 : (!fir.ref<!fir.box<!fir.heap<!fir.array<?x?x!fir.char<1,?>>>>>, i1, !fir.box<!fir.array<?x?x!fir.char<1,?>>>) -> ()
! CHECK: hlfir.end_associate %[[VAL_7]]#1, %[[VAL_7]]#2 : !fir.ref<i32>, i1
! CHECK: return
@@ -113,7 +113,7 @@ subroutine test_char_assumed_size(x)
! CHECK: %[[VAL_20:.*]] = fir.box_addr %[[VAL_5]] : (!fir.box<!fir.array<?x?x!fir.char<1,?>>>) -> !fir.ref<!fir.array<?x?x!fir.char<1,?>>>
! CHECK: %[[VAL_21:.*]] = fir.convert %[[VAL_20]] : (!fir.ref<!fir.array<?x?x!fir.char<1,?>>>) -> !fir.ref<!fir.array<10x?x!fir.char<1,?>>>
! CHECK: %[[VAL_22:.*]] = fir.embox %[[VAL_21]](%[[VAL_18]]) typeparams %[[VAL_19]] : (!fir.ref<!fir.array<10x?x!fir.char<1,?>>>, !fir.shapeshift<2>, index) -> !fir.box<!fir.array<10x?x!fir.char<1,?>>>
-! CHECK: fir.call @takes_char_assumed_size(%[[VAL_22]]) fastmath<contract> {is_bind_c} : (!fir.box<!fir.array<10x?x!fir.char<1,?>>>) -> ()
+! CHECK: fir.call @takes_char_assumed_size(%[[VAL_22]]) proc_attrs<bind_c> fastmath<contract> : (!fir.box<!fir.array<10x?x!fir.char<1,?>>>) -> ()
! CHECK: hlfir.copy_out %[[TMP_BOX]], %[[VAL_2]]#1 to %[[VAL_1]]#0 : (!fir.ref<!fir.box<!fir.heap<!fir.array<?x?x!fir.char<1,?>>>>>, i1, !fir.box<!fir.array<?x?x!fir.char<1,?>>>) -> ()
! CHECK: return
! CHECK: }
@@ -159,7 +159,7 @@ subroutine test_optional_char(x)
! CHECK: %[[VAL_33:.*]] = fir.absent !fir.box<!fir.array<?x!fir.char<1,?>>>
! CHECK: fir.result %[[VAL_33]] : !fir.box<!fir.array<?x!fir.char<1,?>>>
! CHECK: }
-! CHECK: fir.call @takes_optional_char(%[[VAL_15]], %[[VAL_14]]#1) fastmath<contract> {is_bind_c} : (!fir.box<!fir.array<?x!fir.char<1,?>>>, !fir.ref<i32>) -> ()
+! CHECK: fir.call @takes_optional_char(%[[VAL_15]], %[[VAL_14]]#1) proc_attrs<bind_c> fastmath<contract> : (!fir.box<!fir.array<?x!fir.char<1,?>>>, !fir.ref<i32>) -> ()
! CHECK: hlfir.end_associate %[[VAL_14]]#1, %[[VAL_14]]#2 : !fir.ref<i32>, i1
! CHECK: return
! CHECK: }
diff --git a/flang/test/Lower/HLFIR/calls-character-singleton-result.f90 b/flang/test/Lower/HLFIR/calls-character-singleton-result.f90
index b97d9774f734c2..6e0f977862b5a4 100644
--- a/flang/test/Lower/HLFIR/calls-character-singleton-result.f90
+++ b/flang/test/Lower/HLFIR/calls-character-singleton-result.f90
@@ -35,7 +35,7 @@ character(1) function bar() bind(c)
! CHECK: %[[VAL_1:.*]] = fir.alloca !fir.char<1>
! CHECK: %[[VAL_3:.*]] = fir.convert %{{.*}}#0 : (!fir.ref<!fir.char<1,?>>) -> !fir.ref<!fir.char<1>>
! CHECK: %[[VAL_4:.*]]:2 = hlfir.declare %[[VAL_3]] typeparams %{{.*}} {{.*}}Ec
-! CHECK: %[[VAL_5:.*]] = fir.call @bar() fastmath<contract> {is_bind_c} : () -> !fir.char<1>
+! CHECK: %[[VAL_5:.*]] = fir.call @bar() proc_attrs<bind_c> fastmath<contract> : () -> !fir.char<1>
! CHECK: fir.store %[[VAL_5]] to %[[VAL_1]] : !fir.ref<!fir.char<1>>
! CHECK: %[[VAL_6:.*]] = arith.constant false
! CHECK: %[[VAL_7:.*]] = hlfir.as_expr %[[VAL_1]] move %[[VAL_6]] : (!fir.ref<!fir.char<1>>, i1) -> !hlfir.expr<!fir.char<1>>
diff --git a/flang/test/Lower/HLFIR/ignore-type-assumed-shape.f90 b/flang/test/Lower/HLFIR/ignore-type-assumed-shape.f90
index 60951cf1fa73a6..d1c61c6023e089 100644
--- a/flang/test/Lower/HLFIR/ignore-type-assumed-shape.f90
+++ b/flang/test/Lower/HLFIR/ignore-type-assumed-shape.f90
@@ -23,7 +23,7 @@ subroutine test_ignore_t_1(x)
! CHECK: %[[VAL_6:.*]] = fir.shift %[[VAL_5]] : (index) -> !fir.shift<1>
! CHECK: %[[VAL_7:.*]] = fir.rebox %{{.*}}(%[[VAL_6]]) : (!fir.box<!fir.array<10xf32>>, !fir.shift<1>) -> !fir.box<!fir.array<?xf32>>
! CHECK: %[[VAL_8:.*]] = fir.convert %[[VAL_7]] : (!fir.box<!fir.array<?xf32>>) -> !fir.box<!fir.array<?xi32>>
-! CHECK: fir.call @takes_assumed_shape_ignore_tkr_t(%[[VAL_8]]) fastmath<contract> {is_bind_c} : (!fir.box<!fir.array<?xi32>>) -> ()
+! CHECK: fir.call @takes_assumed_shape_ignore_tkr_t(%[[VAL_8]]) proc_attrs<bind_c> fastmath<contract> : (!fir.box<!fir.array<?xi32>>) -> ()
subroutine test_ignore_t_2(x)
use tkr_ifaces
@@ -35,7 +35,7 @@ subroutine test_ignore_t_2(x)
! CHECK: %[[VAL_3:.*]] = fir.shift %[[VAL_2]] : (index) -> !fir.shift<1>
! CHECK: %[[VAL_4:.*]] = fir.rebox %{{.*}}(%[[VAL_3]]) : (!fir.class<!fir.array<?xnone>>, !fir.shift<1>) -> !fir.class<!fir.array<?xnone>>
! CHECK: %[[VAL_5:.*]] = fir.convert %[[VAL_4]] : (!fir.class<!fir.array<?xnone>>) -> !fir.box<!fir.array<?xi32>>
-! CHECK: fir.call @takes_assumed_shape_ignore_tkr_t(%[[VAL_5]]) fastmath<contract> {is_bind_c} : (!fir.box<!fir.array<?xi32>>) -> ()
+! CHECK: fir.call @takes_assumed_shape_ignore_tkr_t(%[[VAL_5]]) proc_attrs<bind_c> fastmath<contract> : (!fir.box<!fir.array<?xi32>>) -> ()
subroutine test_ignore_t_3(x)
use tkr_ifaces
@@ -47,7 +47,7 @@ subroutine test_ignore_t_3(x)
! CHECK: %[[VAL_13:.*]] = fir.shift %[[VAL_12]] : (index) -> !fir.shift<1>
! CHECK: %[[VAL_14:.*]] = fir.rebox %{{.*}}(%[[VAL_13]]) : (!fir.box<!fir.array<10xf32>>, !fir.shift<1>) -> !fir.box<!fir.array<?xf32>>
! CHECK: %[[VAL_15:.*]] = fir.convert %[[VAL_14]] : (!fir.box<!fir.array<?xf32>>) -> !fir.box<!fir.array<?xi32>>
-! CHECK: fir.call @takes_assumed_shape_ignore_tkr_t(%[[VAL_15]]) fastmath<contract> {is_bind_c} : (!fir.box<!fir.array<?xi32>>) -> ()
+! CHECK: fir.call @takes_assumed_shape_ignore_tkr_t(%[[VAL_15]]) proc_attrs<bind_c> fastmath<contract> : (!fir.box<!fir.array<?xi32>>) -> ()
subroutine test_ignore_t_4(x)
use tkr_ifaces
@@ -59,4 +59,4 @@ subroutine test_ignore_t_4(x)
! CHECK: %[[VAL_4:.*]] = fir.shift %[[VAL_3]] : (index) -> !fir.shift<1>
! CHECK: %[[VAL_5:.*]] = fir.rebox %{{.*}}(%[[VAL_4]]) : (!fir.box<!fir.ptr<!fir.array<?xf32>>>, !fir.shift<1>) -> !fir.box<!fir.array<?xf32>>
! CHECK: %[[VAL_6:.*]] = fir.convert %[[VAL_5]] : (!fir.box<!fir.array<?xf32>>) -> !fir.box<!fir.array<?xi32>>
-! CHECK: fir.call @takes_assumed_shape_ignore_tkr_t(%[[VAL_6]]) fastmath<contract> {is_bind_c} : (!fir.box<!fir.array<?xi32>>) -> ()
+! CHECK: fir.call @takes_assumed_shape_ignore_tkr_t(%[[VAL_6]]) proc_attrs<bind_c> fastmath<contract> : (!fir.box<!fir.array<?xi32>>) -> ()
diff --git a/flang/test/Lower/OpenMP/threadprivate-default-clause.f90 b/flang/test/Lower/OpenMP/threadprivate-default-clause.f90
index 65ea30de521678..2a52e1c026a9e5 100644
--- a/flang/test/Lower/OpenMP/threadprivate-default-clause.f90
+++ b/flang/test/Lower/OpenMP/threadprivate-default-clause.f90
@@ -12,7 +12,7 @@
!CHECK: omp.parallel {
!CHECK: %[[A_TP:.*]] = omp.threadprivate %[[A_DECL]]#1 : !fir.ref<i32> -> !fir.ref<i32>
!CHECK: %[[A_TP_DECL:.*]]:2 = hlfir.declare %[[A_TP]] {uniq_name = "_QFsub1Ea"} : (!fir.ref<i32>) -> (!fir.ref<i32>, !fir.ref<i32>)
-!CHECK: %[[TID:.*]] = fir.call @omp_get_thread_num() fastmath<contract> {is_bind_c} : () -> i32
+!CHECK: %[[TID:.*]] = fir.call @omp_get_thread_num() proc_attrs<bind_c> fastmath<contract> : () -> i32
!CHECK: hlfir.assign %[[TID]] to %[[A_TP_DECL]]#0 : i32, !fir.ref<i32>
!CHECK: omp.terminator
!CHECK: }
@@ -46,7 +46,7 @@ subroutine sub1()
!CHECK: %[[A_TP_ADDR:.*]] = fir.coordinate_of %[[BLK_TP_CVT]], %c0_1 : (!fir.ref<!fir.array<?xi8>>, index) -> !fir.ref<i8>
!CHECK: %[[A_TP_ADDR_CVT:.*]] = fir.convert %[[A_TP_ADDR]] : (!fir.ref<i8>) -> !fir.ref<i32>
!CHECK: %[[A_TP_DECL:.*]]:2 = hlfir.declare %[[A_TP_ADDR_CVT]] {uniq_name = "_QFsub2Ea"} : (!fir.ref<i32>) -> (!fir.ref<i32>, !fir.ref<i32>)
-!CHECK: %[[TID:.*]] = fir.call @omp_get_thread_num() fastmath<contract> {is_bind_c} : () -> i32
+!CHECK: %[[TID:.*]] = fir.call @omp_get_thread_num() proc_attrs<bind_c> fastmath<contract> : () -> i32
!CHECK: hlfir.assign %[[TID]] to %[[A_TP_DECL]]#0 : i32, !fir.ref<i32>
!CHECK: omp.terminator
!CHECK: }
diff --git a/flang/test/Lower/block.f90 b/flang/test/Lower/block.f90
index 9d1cb4c298c0aa..70ff67db718edc 100644
--- a/flang/test/Lower/block.f90
+++ b/flang/test/Lower/block.f90
@@ -81,7 +81,7 @@ program bb ! block stack management and exits
! CHECK: %[[V_51:[0-9]+]] = fir.call @llvm.stacksave.p0() fastmath<contract> : () -> !fir.ref<i8>
! CHECK: fir.store %c5{{.*}} to %[[V_0]] : !fir.ref<i32>
- ! CHECK: fir.call @ss(%[[V_0]]) fastmath<contract> {is_bind_c} : (!fir.ref<i32>) -> ()
+ ! CHECK: fir.call @ss(%[[V_0]]) proc_attrs<bind_c> fastmath<contract> : (!fir.ref<i32>) -> ()
! CHECK: fir.call @llvm.stackrestore.p0(%[[V_51]]) fastmath<contract> : (!fir.ref<i8>) -> ()
block
interface
diff --git a/flang/test/Lower/call-bindc.f90 b/flang/test/Lower/call-bindc.f90
index b216a1e4d963d9..f8cafffa3867d8 100644
--- a/flang/test/Lower/call-bindc.f90
+++ b/flang/test/Lower/call-bindc.f90
@@ -18,4 +18,4 @@ program main
end
! CHECK-LABEL: func.func @_QQmain()
-! CHECK: fir.call %{{.*}}(%{{.*}}) fastmath<contract> {is_bind_c} : (i32) -> !fir.complex<4>
+! CHECK: fir.call %{{.*}}(%{{.*}}) proc_attrs<bind_c> fastmath<contract> : (i32) -> !fir.complex<4>
>From 57b03732eacbb253c9b1972113ee7bb5eb467d22 Mon Sep 17 00:00:00 2001
From: Jean Perier <jperier at nvidia.com>
Date: Fri, 23 Aug 2024 00:11:51 -0700
Subject: [PATCH 2/2] remove useless getMnemonic() line
---
flang/lib/Optimizer/Dialect/FIROps.cpp | 1 -
1 file changed, 1 deletion(-)
diff --git a/flang/lib/Optimizer/Dialect/FIROps.cpp b/flang/lib/Optimizer/Dialect/FIROps.cpp
index c2a893aee7b0a5..ce0ae446bb7af8 100644
--- a/flang/lib/Optimizer/Dialect/FIROps.cpp
+++ b/flang/lib/Optimizer/Dialect/FIROps.cpp
@@ -1108,7 +1108,6 @@ void fir::CallOp::print(mlir::OpAsmPrinter &p) {
if (procAttrs &&
procAttrs.getValue() != fir::FortranProcedureFlagsEnum::none) {
p << ' ' << fir::FortranProcedureFlagsEnumAttr::getMnemonic();
- mlir::arith::FastMathFlagsAttr::getMnemonic();
p.printStrippedAttrOrType(procAttrs);
}
More information about the flang-commits
mailing list