Issue 204586
Summary [mlir](convert-gpu-to-nvvm) mlir-opt crashes at BuiltinAttributes.cpp:127 with assertion `!findDuplicateElement(array) && "DictionaryAttr element names must be unique"' failed.
Labels mlir
Assignees
Reporter Aran-414
    git version: 08b557bf5139394194be22359268b7b88f81faae

MLIR Program:
```mlir
"builtin.module"() ({
  "gpu.module"() <{sym_name = "m1"}> ({
    "gpu.func"() <{function_type = () -> ()}> ({
 "gpu.return"() : () -> ()
    }) {gpu.kernel, spirv.entry_point_abi = #spirv.entry_point_abi<workgroup_size = [16, 1, 1]>, sym_name = "builtin_global_id_y", workgroup_attributions = 0 : i64} : () -> ()
  }) : () -> ()
}) {gpu.container_module, spirv.target_env = #spirv.target_env<#spirv.vce<v1.0, [Shader, Int64], []>, #spirv.resource_limits<>>} : () -> ()
```

Reproduce Command: `mlir-opt -convert-gpu-to-nvvm a.mlir`

StackTrace:
```
mlir-opt: /data2/dependency/dev/llvm-project/mlir/lib/IR/BuiltinAttributes.cpp:127: static bool mlir::DictionaryAttr::sortInPlace(llvm::SmallVectorImpl<mlir::NamedAttribute>&): Assertion `!findDuplicateElement(array) && "DictionaryAttr element names must be unique"' failed.
PLEASE submit a bug report to https://github.com/llvm/llvm-project/issues/ and include the crash backtrace and instructions to reproduce the bug.
Stack dump:
0.      Program arguments: /data2/dependency/dev/llvm-project/build/bin/mlir-opt -convert-gpu-to-nvvm a.mlir
 #0 0x000056312a1b9f6f llvm::sys::PrintStackTrace(llvm::raw_ostream&, int) (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0x25bbf6f)
 #1 0x000056312a1b673c llvm::sys::RunSignalHandlers() (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0x25b873c)
 #2 0x000056312a1b68d7 SignalHandler(int, siginfo_t*, void*) Signals.cpp:0:0
 #3 0x00007f5d56603420 __restore_rt (/lib/x86_64-linux-gnu/libpthread.so.0+0x14420)
 #4 0x00007f5d560ee00b raise (/lib/x86_64-linux-gnu/libc.so.6+0x4300b)
 #5 0x00007f5d560cd859 abort (/lib/x86_64-linux-gnu/libc.so.6+0x22859)
 #6 0x00007f5d560cd729 (/lib/x86_64-linux-gnu/libc.so.6+0x22729)
 #7 0x00007f5d560defd6 (/lib/x86_64-linux-gnu/libc.so.6+0x33fd6)
 #8 0x0000563133280a05 mlir::DictionaryAttr::sortInPlace(llvm::SmallVectorImpl<mlir::NamedAttribute>&) (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0xb682a05)
 #9 0x0000563133348275 mlir::NamedAttrList::getDictionary(mlir::MLIRContext*) const (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0xb74a275)
#10 0x000056313333bc7b mlir::Operation::create(mlir::OperationState const&) (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0xb73dc7b)
#11 0x0000563133275ec3 mlir::OpBuilder::create(mlir::OperationState const&) (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0xb677ec3)
#12 0x00005631325aab9a mlir::LLVM::LLVMFuncOp::create(mlir::OpBuilder&, mlir::Location, mlir::LLVM::detail::LLVMFuncOpGenericAdaptorBase::Properties const&, llvm::ArrayRef<mlir::NamedAttribute>) (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0xa9acb9a)
#13 0x000056312bd53bc0 mlir::GPUFuncOpLowering::matchAndRewrite(mlir::gpu::GPUFuncOp, mlir::gpu::GPUFuncOpAdaptor, mlir::ConversionPatternRewriter&) const (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0x4155bc0)
#14 0x000056312bb19ebe mlir::ConvertOpToLLVMPattern<mlir::gpu::GPUFuncOp, false>::matchAndRewrite(mlir::Operation*, llvm::ArrayRef<mlir::ValueRange>, mlir::ConversionPatternRewriter&) const (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0x3f1bebe)
#15 0x0000563132f0de58 mlir::ConversionPattern::matchAndRewrite(mlir::Operation*, mlir::PatternRewriter&) const (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0xb30fe58)
#16 0x0000563132f54aa2 mlir::PatternApplicator::matchAndRewrite(mlir::Operation*, mlir::PatternRewriter&, llvm::function_ref<bool (mlir::Pattern const&)>, llvm::function_ref<void (mlir::Pattern const&)>, llvm::function_ref<llvm::LogicalResult (mlir::Pattern const&)>)::'lambda'()::operator()() const PatternApplicator.cpp:0:0
#17 0x0000563132f55f98 mlir::PatternApplicator::matchAndRewrite(mlir::Operation*, mlir::PatternRewriter&, llvm::function_ref<bool (mlir::Pattern const&)>, llvm::function_ref<void (mlir::Pattern const&)>, llvm::function_ref<llvm::LogicalResult (mlir::Pattern const&)>) (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0xb357f98)
#18 0x0000563132f08aeb (anonymous namespace)::OperationLegalizer::legalize(mlir::Operation*) DialectConversion.cpp:0:0
#19 0x0000563132f08fe6 mlir::OperationConverter::convert(mlir::Operation*, bool) (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0xb30afe6)
#20 0x0000563132f13f1a mlir::OperationConverter::applyConversion(llvm::ArrayRef<mlir::Operation*>) (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0xb315f1a)
#21 0x0000563132f157f8 applyConversion(llvm::ArrayRef<mlir::Operation*>, mlir::ConversionTarget const&, mlir::FrozenRewritePatternSet const&, mlir::ConversionConfig, (anonymous namespace)::OpConversionMode) DialectConversion.cpp:0:0
#22 0x0000563132f159d6 mlir::applyPartialConversion(mlir::Operation*, mlir::ConversionTarget const&, mlir::FrozenRewritePatternSet const&, mlir::ConversionConfig) (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0xb3179d6)
#23 0x000056312bb2b79a (anonymous namespace)::LowerGpuOpsToNVVMOpsPass::runOnOperation() LowerGpuOpsToNVVMOps.cpp:0:0
#24 0x0000563132fc0bf1 mlir::detail::OpToOpPassAdaptor::run(mlir::Pass*, mlir::Operation*, mlir::AnalysisManager, bool, unsigned int) (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0xb3c2bf1)
#25 0x0000563132fc0f80 mlir::detail::OpToOpPassAdaptor::runPipeline(mlir::OpPassManager&, mlir::Operation*, mlir::AnalysisManager, bool, unsigned int, mlir::PassInstrumentor*, mlir::PassInstrumentation::PipelineParentInfo const*) (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0xb3c2f80)
#26 0x0000563132fc185e mlir::detail::OpToOpPassAdaptor::runOnOperationAsyncImpl(bool)::'lambda12'(mlir::detail::OpToOpPassAdaptor::runOnOperationAsyncImpl(bool)::OpPMInfo&)::operator()(mlir::detail::OpToOpPassAdaptor::runOnOperationAsyncImpl(bool)::OpPMInfo&) const Pass.cpp:0:0
#27 0x0000563132fbfba0 mlir::detail::OpToOpPassAdaptor::runOnOperationAsyncImpl(bool) (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0xb3c1ba0)
#28 0x0000563132fc0acb mlir::detail::OpToOpPassAdaptor::run(mlir::Pass*, mlir::Operation*, mlir::AnalysisManager, bool, unsigned int) (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0xb3c2acb)
#29 0x0000563132fc0f80 mlir::detail::OpToOpPassAdaptor::runPipeline(mlir::OpPassManager&, mlir::Operation*, mlir::AnalysisManager, bool, unsigned int, mlir::PassInstrumentor*, mlir::PassInstrumentation::PipelineParentInfo const*) (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0xb3c2f80)
#30 0x0000563132fc3034 mlir::PassManager::runPasses(mlir::Operation*, mlir::AnalysisManager) (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0xb3c5034)
#31 0x0000563132fc40c1 mlir::PassManager::run(mlir::Operation*) (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0xb3c60c1)
#32 0x000056312a27f86b performActions(llvm::raw_ostream&, std::shared_ptr<llvm::SourceMgr> const&, mlir::MLIRContext*, mlir::MlirOptMainConfig const&) MlirOptMain.cpp:0:0
#33 0x000056312a280313 processBuffer(llvm::raw_ostream&, std::unique_ptr<llvm::MemoryBuffer, std::default_delete<llvm::MemoryBuffer>>, llvm::MemoryBufferRef, mlir::MlirOptMainConfig const&, mlir::DialectRegistry&, mlir::SourceMgrDiagnosticVerifierHandler*, llvm::ThreadPoolInterface*) MlirOptMain.cpp:0:0
#34 0x000056312a28059b llvm::LogicalResult llvm::function_ref<llvm::LogicalResult (std::unique_ptr<llvm::MemoryBuffer, std::default_delete<llvm::MemoryBuffer>>, llvm::MemoryBufferRef const&, llvm::raw_ostream&)>::callback_fn<mlir::MlirOptMain(llvm::raw_ostream&, std::unique_ptr<llvm::MemoryBuffer, std::default_delete<llvm::MemoryBuffer>>, mlir::DialectRegistry&, mlir::MlirOptMainConfig const&)::'lambda'(std::unique_ptr<llvm::MemoryBuffer, std::default_delete<llvm::MemoryBuffer>>, llvm::MemoryBufferRef, llvm::raw_ostream&)>(long, std::unique_ptr<llvm::MemoryBuffer, std::default_delete<llvm::MemoryBuffer>>, llvm::MemoryBufferRef const&, llvm::raw_ostream&) MlirOptMain.cpp:0:0
#35 0x000056313339bb45 mlir::splitAndProcessBuffer(std::unique_ptr<llvm::MemoryBuffer, std::default_delete<llvm::MemoryBuffer>>, llvm::function_ref<llvm::LogicalResult (std::unique_ptr<llvm::MemoryBuffer, std::default_delete<llvm::MemoryBuffer>>, llvm::MemoryBufferRef const&, llvm::raw_ostream&)>, llvm::raw_ostream&, llvm::StringRef, llvm::StringRef) (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0xb79db45)
#36 0x000056312a2766a4 mlir::MlirOptMain(llvm::raw_ostream&, std::unique_ptr<llvm::MemoryBuffer, std::default_delete<llvm::MemoryBuffer>>, mlir::DialectRegistry&, mlir::MlirOptMainConfig const&) (.part.0) MlirOptMain.cpp:0:0
#37 0x000056312a280ab7 mlir::MlirOptMain(int, char**, llvm::StringRef, llvm::StringRef, mlir::DialectRegistry&) (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0x2682ab7)
#38 0x000056312a280cf3 mlir::MlirOptMain(int, char**, llvm::StringRef, mlir::DialectRegistry&) (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0x2682cf3)
#39 0x000056312a0e6aec main (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0x24e8aec)
#40 0x00007f5d560cf083 __libc_start_main (/lib/x86_64-linux-gnu/libc.so.6+0x24083)
#41 0x000056312a19667e _start (/data2/dependency/dev/llvm-project/build/bin/mlir-opt+0x259867e)
Aborted (core dumped)
```
_______________________________________________
llvm-bugs mailing list
[email protected]
https://lists.llvm.org/cgi-bin/mailman/listinfo/llvm-bugs

Reply via email to