Skip to content

Commit 2051a7b

Browse files
jeanPerierjeanPerier
authored
[flang][NFC] turn fir.call is_bind_c into enum for procedure flags (#105691)
First patch to fix a BIND(C) ABI issue (#102113). I need to keep track of BIND(C) in more locations (fir.dispatch and func.func operations), 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 changes low here. Adding the attribute on fir.dispatch and func.func will also be done in separate patches.
1 parent 2f144ac commit 2051a7b

22 files changed

+100
-48
lines changed

‎flang/include/flang/Optimizer/Dialect/FIRAttr.td

Lines changed: 27 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -58,6 +58,33 @@ def fir_FortranVariableFlagsAttr : fir_Attr<"FortranVariableFlags"> {
5858
"::fir::FortranVariableFlagsAttr::get($_builder.getContext(), $0)";
5959
}
6060

61+
62+
/// Fortran procedure attributes (F2023 15.6.2.1). BIND attribute (18.3.7)
63+
/// is also tracked in the same enum. Recursive (resp. Impure) attribute
64+
/// is implied by the absence of opposite NonRecursive (resp. Pure) attribute.
65+
def FIRfuncNoAttributes : I32BitEnumAttrCaseNone<"none">;
66+
def FIRfuncElemental : I32BitEnumAttrCaseBit<"elemental", 0>;
67+
def FIRfuncPure : I32BitEnumAttrCaseBit<"pure", 1>;
68+
def FIRfuncNonRecursive : I32BitEnumAttrCaseBit<"non_recursive", 2>;
69+
def FIRfuncSimple : I32BitEnumAttrCaseBit<"simple", 3>;
70+
def FIRfuncBind_c : I32BitEnumAttrCaseBit<"bind_c", 4>;
71+
72+
def fir_FortranProcedureFlagsEnum : I32BitEnumAttr<
73+
"FortranProcedureFlagsEnum",
74+
"Fortran procedure attributes",
75+
[FIRfuncNoAttributes, FIRfuncElemental, FIRfuncPure, FIRfuncNonRecursive,
76+
FIRfuncSimple, FIRfuncBind_c]> {
77+
let separator = ", ";
78+
let cppNamespace = "::fir";
79+
let genSpecializedAttr = 0;
80+
let printBitEnumPrimaryGroups = 1;
81+
}
82+
83+
def fir_FortranProcedureFlagsAttr :
84+
EnumAttr<FIROpsDialect, fir_FortranProcedureFlagsEnum, "proc_attrs"> {
85+
let assemblyFormat = "`<` $value `>`";
86+
}
87+
6188
def fir_BoxFieldAttr : I32EnumAttr<
6289
"BoxFieldAttr", "",
6390
[

‎flang/include/flang/Optimizer/Dialect/FIROps.td

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -2432,9 +2432,9 @@ def fir_CallOp : fir_Op<"call",
24322432
let arguments = (ins
24332433
OptionalAttr<SymbolRefAttr>:$callee,
24342434
Variadic<AnyType>:$args,
2435+
OptionalAttr<fir_FortranProcedureFlagsAttr>:$procedure_attrs,
24352436
DefaultValuedAttr<Arith_FastMathAttr,
2436-
"::mlir::arith::FastMathFlags::none">:$fastmath,
2437-
UnitAttr:$is_bind_c
2437+
"::mlir::arith::FastMathFlags::none">:$fastmath
24382438
);
24392439
let results = (outs Variadic<AnyType>);
24402440

‎flang/lib/Lower/ConvertCall.cpp

Lines changed: 11 additions & 7 deletions
Original file line numberDiff line numberDiff line change
@@ -372,17 +372,17 @@ std::pair<fir::ExtendedValue, bool> Fortran::lower::genCallOpAndResult(
372372
auto stackSaveSymbol = bldr->getSymbolRefAttr(stackSaveFn.getName());
373373
mlir::Value sp;
374374
fir::CallOp call = bldr->create<fir::CallOp>(
375-
loc, stackSaveFn.getFunctionType().getResults(), stackSaveSymbol,
375+
loc, stackSaveSymbol, stackSaveFn.getFunctionType().getResults(),
376376
mlir::ValueRange{});
377377
if (call.getNumResults() != 0)
378378
sp = call.getResult(0);
379379
stmtCtx.attachCleanup([bldr, loc, sp]() {
380380
auto stackRestoreFn = fir::factory::getLlvmStackRestore(*bldr);
381381
auto stackRestoreSymbol =
382382
bldr->getSymbolRefAttr(stackRestoreFn.getName());
383-
bldr->create<fir::CallOp>(loc,
383+
bldr->create<fir::CallOp>(loc, stackRestoreSymbol,
384384
stackRestoreFn.getFunctionType().getResults(),
385-
stackRestoreSymbol, mlir::ValueRange{sp});
385+
mlir::ValueRange{sp});
386386
});
387387
}
388388
mlir::Value temp =
@@ -640,11 +640,15 @@ std::pair<fir::ExtendedValue, bool> Fortran::lower::genCallOpAndResult(
640640
if (callNumResults != 0)
641641
callResult = dispatch.getResult(0);
642642
} else {
643-
// Standard procedure call with fir.call.
644-
auto call = builder.create<fir::CallOp>(loc, funcType.getResults(),
645-
funcSymbolAttr, operands);
643+
// TODO: gather other procedure attributes.
644+
fir::FortranProcedureFlagsEnumAttr procAttrs;
646645
if (caller.characterize().IsBindC())
647-
call.setIsBindC(true);
646+
procAttrs = fir::FortranProcedureFlagsEnumAttr::get(
647+
builder.getContext(), fir::FortranProcedureFlagsEnum::bind_c);
648+
649+
// Standard procedure call with fir.call.
650+
auto call = builder.create<fir::CallOp>(
651+
loc, funcType.getResults(), funcSymbolAttr, operands, procAttrs);
648652

649653
callNumResults = call.getNumResults();
650654
if (callNumResults != 0)

‎flang/lib/Lower/ConvertExpr.cpp

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -6120,7 +6120,7 @@ class ArrayExprLowering {
61206120
mlir::SymbolRefAttr funcSymAttr =
61216121
builder.getSymbolRefAttr(memcpyFunc.getName());
61226122
mlir::FunctionType funcTy = memcpyFunc.getFunctionType();
6123-
builder.create<fir::CallOp>(loc, funcTy.getResults(), funcSymAttr, args);
6123+
builder.create<fir::CallOp>(loc, funcSymAttr, funcTy.getResults(), args);
61246124
}
61256125

61266126
// Construct code to check for a buffer overrun and realloc the buffer when
@@ -6146,7 +6146,7 @@ class ArrayExprLowering {
61466146
builder.getSymbolRefAttr(reallocFunc.getName());
61476147
mlir::FunctionType funcTy = reallocFunc.getFunctionType();
61486148
auto newMem = builder.create<fir::CallOp>(
6149-
loc, funcTy.getResults(), funcSymAttr,
6149+
loc, funcSymAttr, funcTy.getResults(),
61506150
llvm::ArrayRef<mlir::Value>{
61516151
builder.createConvert(loc, funcTy.getInputs()[0], mem),
61526152
builder.createConvert(loc, funcTy.getInputs()[1], byteSz)});

‎flang/lib/Optimizer/Builder/IntrinsicCall.cpp

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -819,8 +819,8 @@ mlir::Value genLibCall(fir::FirOpBuilder &builder, mlir::Location loc,
819819

820820
llvm::SmallVector<mlir::Value, 3> operands{funcPointer};
821821
operands.append(args.begin(), args.end());
822-
libCall = builder.create<fir::CallOp>(loc, libFuncType.getResults(),
823-
nullptr, operands);
822+
libCall = builder.create<fir::CallOp>(loc, mlir::SymbolRefAttr{},
823+
libFuncType.getResults(), operands);
824824
}
825825

826826
LLVM_DEBUG(libCall.dump(); llvm::dbgs() << "\n");

‎flang/lib/Optimizer/Dialect/FIRAttr.cpp

Lines changed: 2 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -296,7 +296,8 @@ void fir::printFirAttribute(FIROpsDialect *dialect, mlir::Attribute attr,
296296
//===----------------------------------------------------------------------===//
297297

298298
void FIROpsDialect::registerAttributes() {
299-
addAttributes<ClosedIntervalAttr, ExactTypeAttr, FortranVariableFlagsAttr,
299+
addAttributes<ClosedIntervalAttr, ExactTypeAttr,
300+
FortranProcedureFlagsEnumAttr, FortranVariableFlagsAttr,
300301
LowerBoundAttr, PointIntervalAttr, RealAttr, ReduceAttr,
301302
SubclassAttr, UpperBoundAttr, LocationKindAttr,
302303
LocationKindArrayAttr>();

‎flang/lib/Optimizer/Dialect/FIROps.cpp

Lines changed: 20 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -1103,6 +1103,14 @@ void fir::CallOp::print(mlir::OpAsmPrinter &p) {
11031103
p << getOperand(0);
11041104
p << '(' << (*this)->getOperands().drop_front(isDirect ? 0 : 1) << ')';
11051105

1106+
// Print `proc_attrs<...>`, if present.
1107+
fir::FortranProcedureFlagsEnumAttr procAttrs = getProcedureAttrsAttr();
1108+
if (procAttrs &&
1109+
procAttrs.getValue() != fir::FortranProcedureFlagsEnum::none) {
1110+
p << ' ' << fir::FortranProcedureFlagsEnumAttr::getMnemonic();
1111+
p.printStrippedAttrOrType(procAttrs);
1112+
}
1113+
11061114
// Print 'fastmath<...>' (if it has non-default value) before
11071115
// any other attributes.
11081116
mlir::arith::FastMathFlagsAttr fmfAttr = getFastmathAttr();
@@ -1111,9 +1119,9 @@ void fir::CallOp::print(mlir::OpAsmPrinter &p) {
11111119
p.printStrippedAttrOrType(fmfAttr);
11121120
}
11131121

1114-
p.printOptionalAttrDict(
1115-
(*this)->getAttrs(),
1116-
{fir::CallOp::getCalleeAttrNameStr(), getFastmathAttrName()});
1122+
p.printOptionalAttrDict((*this)->getAttrs(),
1123+
{fir::CallOp::getCalleeAttrNameStr(),
1124+
getFastmathAttrName(), getProcedureAttrsAttrName()});
11171125
auto resultTypes{getResultTypes()};
11181126
llvm::SmallVector<mlir::Type> argTypes(
11191127
llvm::drop_begin(getOperandTypes(), isDirect ? 0 : 1));
@@ -1138,6 +1146,15 @@ mlir::ParseResult fir::CallOp::parse(mlir::OpAsmParser &parser,
11381146
if (parser.parseOperandList(operands, mlir::OpAsmParser::Delimiter::Paren))
11391147
return mlir::failure();
11401148

1149+
// Parse `proc_attrs<...>`, if present.
1150+
fir::FortranProcedureFlagsEnumAttr procAttr;
1151+
if (mlir::succeeded(parser.parseOptionalKeyword(
1152+
fir::FortranProcedureFlagsEnumAttr::getMnemonic())))
1153+
if (parser.parseCustomAttributeWithFallback(
1154+
procAttr, mlir::Type{}, getProcedureAttrsAttrName(result.name),
1155+
attrs))
1156+
return mlir::failure();
1157+
11411158
// Parse 'fastmath<...>', if present.
11421159
mlir::arith::FastMathFlagsAttr fmfAttr;
11431160
llvm::StringRef fmfAttrName = getFastmathAttrName(result.name);

‎flang/lib/Optimizer/Transforms/ConstantArgumentGlobalisation.cpp

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -126,10 +126,10 @@ class CallOpRewriter : public mlir::OpRewritePattern<fir::CallOp> {
126126
newResultTypes.append(callOp.getResultTypes().begin(),
127127
callOp.getResultTypes().end());
128128
fir::CallOp newOp = builder.create<fir::CallOp>(
129-
loc, newResultTypes,
129+
loc,
130130
callOp.getCallee().has_value() ? callOp.getCallee().value()
131131
: mlir::SymbolRefAttr{},
132-
newOperands);
132+
newResultTypes, newOperands);
133133
// Copy all the attributes from the old to new op.
134134
newOp->setAttrs(callOp->getAttrs());
135135
rewriter.replaceOp(callOp, newOp);

‎flang/lib/Optimizer/Transforms/PolymorphicOpConversion.cpp

Lines changed: 4 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -205,7 +205,10 @@ struct DispatchOpConv : public OpConversionPattern<fir::DispatchOp> {
205205
// Make the call.
206206
llvm::SmallVector<mlir::Value> args{funcPtr};
207207
args.append(dispatch.getArgs().begin(), dispatch.getArgs().end());
208-
rewriter.replaceOpWithNewOp<fir::CallOp>(dispatch, resTypes, nullptr, args);
208+
// FIXME: add procedure_attrs to fir.dispatch and propagate to fir.call.
209+
rewriter.replaceOpWithNewOp<fir::CallOp>(
210+
dispatch, resTypes, nullptr, args,
211+
/*procedure_attrs=*/fir::FortranProcedureFlagsEnumAttr{});
209212
return mlir::success();
210213
}
211214

‎flang/lib/Optimizer/Transforms/StackArrays.cpp

Lines changed: 4 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -741,9 +741,9 @@ void AllocMemConversion::insertStackSaveRestore(
741741
builder.setInsertionPoint(oldAlloc);
742742
mlir::Value sp =
743743
builder
744-
.create<fir::CallOp>(oldAlloc.getLoc(),
744+
.create<fir::CallOp>(oldAlloc.getLoc(), stackSaveSym,
745745
stackSaveFn.getFunctionType().getResults(),
746-
stackSaveSym, mlir::ValueRange{})
746+
mlir::ValueRange{})
747747
.getResult(0);
748748

749749
mlir::func::FuncOp stackRestoreFn =
@@ -753,9 +753,9 @@ void AllocMemConversion::insertStackSaveRestore(
753753

754754
auto createStackRestoreCall = [&](mlir::Operation *user) {
755755
builder.setInsertionPoint(user);
756-
builder.create<fir::CallOp>(user->getLoc(),
756+
builder.create<fir::CallOp>(user->getLoc(), stackRestoreSym,
757757
stackRestoreFn.getFunctionType().getResults(),
758-
stackRestoreSym, mlir::ValueRange{sp});
758+
mlir::ValueRange{sp});
759759
};
760760

761761
for (mlir::Operation *user : oldAlloc->getUsers()) {

‎flang/test/HLFIR/c_ptr_byvalue.f90

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -7,7 +7,7 @@
77
! CHECK: %[[VAL_113:.*]] = fir.load %[[VAL_112]] : !fir.ref<i64>
88
! CHECK: %[[VAL_114:.*]] = fir.convert %[[VAL_113]] : (i64) -> !fir.ref<i64>
99
! CHECK: hlfir.end_associate %[[VAL_110]]#1, %[[VAL_110]]#2 : !fir.ref<!fir.type<_QM__fortran_builtinsT__builtin_c_ptr{__address:i64}>>, i1
10-
! CHECK: fir.call @get_expected_f(%[[VAL_114]]) fastmath<contract> {is_bind_c} : (!fir.ref<i64>) -> ()
10+
! CHECK: fir.call @get_expected_f(%[[VAL_114]]) proc_attrs<bind_c> fastmath<contract> : (!fir.ref<i64>) -> ()
1111
subroutine test1
1212
use iso_c_binding
1313
interface
@@ -28,7 +28,7 @@ end subroutine get_expected_f
2828
! 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>
2929
! CHECK: %[[VAL_100:.*]] = fir.load %[[VAL_99]] : !fir.ref<i64>
3030
! CHECK: %[[VAL_101:.*]] = fir.convert %[[VAL_100]] : (i64) -> !fir.ref<i64>
31-
! CHECK: fir.call @get_expected_f(%[[VAL_101]]) fastmath<contract> {is_bind_c} : (!fir.ref<i64>) -> ()
31+
! CHECK: fir.call @get_expected_f(%[[VAL_101]]) proc_attrs<bind_c> fastmath<contract> : (!fir.ref<i64>) -> ()
3232
subroutine test2(cptr)
3333
use iso_c_binding
3434
interface

‎flang/test/Lower/CUDA/cuda-device-proc.cuf

Lines changed: 4 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -18,13 +18,13 @@ end
1818

1919
! CHECK-LABEL: func.func @_QPdevsub() attributes {cuf.proc_attr = #cuf.cuda_proc<global>}
2020
! CHECK: fir.call @__syncthreads()
21-
! CHECK: fir.call @__syncwarp(%{{.*}}) fastmath<contract> {is_bind_c} : (!fir.ref<i32>) -> ()
21+
! CHECK: fir.call @__syncwarp(%{{.*}}) proc_attrs<bind_c> fastmath<contract> : (!fir.ref<i32>) -> ()
2222
! CHECK: fir.call @__threadfence()
2323
! CHECK: fir.call @__threadfence_block()
2424
! CHECK: fir.call @__threadfence_system()
25-
! CHECK: %{{.*}} = fir.call @__syncthreads_and(%{{.*}}) fastmath<contract> {is_bind_c} : (!fir.ref<i32>) -> i32
26-
! CHECK: %{{.*}} = fir.call @__syncthreads_count(%{{.*}}) fastmath<contract> {is_bind_c} : (!fir.ref<i32>) -> i32
27-
! CHECK: %{{.*}} = fir.call @__syncthreads_or(%{{.*}}) fastmath<contract> {is_bind_c} : (!fir.ref<i32>) -> i32
25+
! CHECK: %{{.*}} = fir.call @__syncthreads_and(%{{.*}}) proc_attrs<bind_c> fastmath<contract> : (!fir.ref<i32>) -> i32
26+
! CHECK: %{{.*}} = fir.call @__syncthreads_count(%{{.*}}) proc_attrs<bind_c> fastmath<contract> : (!fir.ref<i32>) -> i32
27+
! CHECK: %{{.*}} = fir.call @__syncthreads_or(%{{.*}}) proc_attrs<bind_c> fastmath<contract> : (!fir.ref<i32>) -> i32
2828

2929
! CHECK: func.func private @__syncthreads() attributes {cuf.proc_attr = #cuf.cuda_proc<device>, fir.bindc_name = "__syncthreads"}
3030
! 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"}

‎flang/test/Lower/HLFIR/assumed-rank-calls.f90

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -36,7 +36,7 @@ subroutine bindc_func(x) bind(c)
3636
! CHECK: %[[VAL_1:.*]] = fir.dummy_scope : !fir.dscope
3737
! 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>>)
3838
! CHECK: %[[VAL_3:.*]] = fir.rebox_assumed_rank %[[VAL_2]]#0 lbs zeroes : (!fir.box<!fir.array<*:f32>>) -> !fir.box<!fir.array<*:f32>>
39-
! CHECK: fir.call @bindc_func(%[[VAL_3]]) fastmath<contract> {is_bind_c} : (!fir.box<!fir.array<*:f32>>) -> ()
39+
! CHECK: fir.call @bindc_func(%[[VAL_3]]) proc_attrs<bind_c> fastmath<contract> : (!fir.box<!fir.array<*:f32>>) -> ()
4040
! CHECK: return
4141
! CHECK: }
4242

‎flang/test/Lower/HLFIR/assumed-rank-iface.f90

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -38,7 +38,7 @@ subroutine int_scalar_to_assumed_rank_bindc(x)
3838
! 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>)
3939
! CHECK: %[[VAL_2:.*]] = fir.embox %[[VAL_1]]#0 : (!fir.ref<i32>) -> !fir.box<i32>
4040
! CHECK: %[[VAL_3:.*]] = fir.convert %[[VAL_2]] : (!fir.box<i32>) -> !fir.box<!fir.array<*:i32>>
41-
! CHECK: fir.call @int_assumed_rank_bindc(%[[VAL_3]]) fastmath<contract> {is_bind_c} : (!fir.box<!fir.array<*:i32>>) -> ()
41+
! CHECK: fir.call @int_assumed_rank_bindc(%[[VAL_3]]) proc_attrs<bind_c> fastmath<contract> : (!fir.box<!fir.array<*:i32>>) -> ()
4242

4343
subroutine int_r1_to_assumed_rank(x)
4444
use ifaces, only : int_assumed_rank
@@ -94,7 +94,7 @@ subroutine int_assumed_shape_to_assumed_rank_bindc(x)
9494
! CHECK: %[[VAL_3:.*]] = fir.shift %[[VAL_2]], %[[VAL_2]] : (index, index) -> !fir.shift<2>
9595
! 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>>
9696
! CHECK: %[[VAL_5:.*]] = fir.convert %[[VAL_4]] : (!fir.box<!fir.array<?x?xi32>>) -> !fir.box<!fir.array<*:i32>>
97-
! CHECK: fir.call @int_assumed_rank_bindc(%[[VAL_5]]) fastmath<contract> {is_bind_c} : (!fir.box<!fir.array<*:i32>>) -> ()
97+
! CHECK: fir.call @int_assumed_rank_bindc(%[[VAL_5]]) proc_attrs<bind_c> fastmath<contract> : (!fir.box<!fir.array<*:i32>>) -> ()
9898

9999
subroutine int_allocatable_to_assumed_rank(x)
100100
use ifaces, only : int_assumed_rank

‎flang/test/Lower/HLFIR/bindc-value-derived.f90

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -31,7 +31,7 @@ subroutine call_it(x)
3131
! CHECK-SAME: %[[VAL_0:.*]]: !fir.ref<!fir.type<_QMbindc_byvalTt{i:i32}>> {fir.bindc_name = "x"}) {
3232
! 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}>>)
3333
! CHECK: %[[VAL_2:.*]] = fir.load %[[VAL_1]]#1 : !fir.ref<!fir.type<_QMbindc_byvalTt{i:i32}>>
34-
! CHECK: fir.call @test(%[[VAL_2]]) fastmath<contract> {is_bind_c} : (!fir.type<_QMbindc_byvalTt{i:i32}>) -> ()
34+
! CHECK: fir.call @test(%[[VAL_2]]) proc_attrs<bind_c> fastmath<contract> : (!fir.type<_QMbindc_byvalTt{i:i32}>) -> ()
3535
! CHECK: return
3636
! CHECK: }
3737
end module

‎flang/test/Lower/HLFIR/block_bindc_pocs.f90

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -9,7 +9,7 @@ end subroutine test_proc
99
end interface
1010
end module m
1111
!CHECK-DAG: %[[S0:.*]] = fir.call @llvm.stacksave.p0() fastmath<contract> : () -> !fir.ref<i8>
12-
!CHECK-DAG: fir.call @test_proc() fastmath<contract> {is_bind_c} : () -> ()
12+
!CHECK-DAG: fir.call @test_proc() proc_attrs<bind_c> fastmath<contract> : () -> ()
1313
!CHECK-DAG: fir.call @llvm.stackrestore.p0(%[[S0]]) fastmath<contract> : (!fir.ref<i8>) -> ()
1414
!CHECK-DAG: func.func private @test_proc() attributes {fir.bindc_name = "test_proc"}
1515
subroutine test

0 commit comments

Comments
 (0)