[flang][NFC] turn fir.call is_bind_c into enum for procedure flags #105691
Conversation
This file contains 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
…lags First patch to fix a BIND(C) ABI issue (llvm#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.
llvmbot
added
flang
|
@llvm/pr-subscribers-flang-openmp @llvm/pr-subscribers-flang-fir-hlfir Author: None (jeanPerier) ChangesFirst 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. Patch is 31.84 KiB, truncated to 20.00 KiB below, full version: https://github.com/llvm/llvm-project/pull/105691.diff 22 Files Affected:
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/HL...
[truncated]
|
vzakhari
approved these changes
Aug 22, 2024
psteinfeld
approved these changes
Aug 22, 2024
cjdb
pushed a commit
to cjdb/llvm-project
that referenced
this pull request
Aug 23, 2024
…lvm#105691) First patch to fix a BIND(C) ABI issue (llvm#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.
5chmidti
pushed a commit
that referenced
this pull request
Aug 24, 2024
…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.
dmpolukhin
pushed a commit
to dmpolukhin/llvm-project
that referenced
this pull request
Sep 2, 2024
…lvm#105691) First patch to fix a BIND(C) ABI issue (llvm#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.
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.
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.