| Issue |
173378
|
| Summary |
mlir-opt crashes in mlir/include/mlir/Dialect/NVGPU/IR/NVGPUOps.h.inc:3612
|
| Labels |
crash,
mlir:nvgpu
|
| Assignees |
|
| Reporter |
Emilyaxe
|
When running opt on the following IR, mlir-opt crashes.
version 2c02e4c7909a
test.mlir
```
func.func @matmul_scalar(%arg0: vector<2x1xf32>, %arg1: vector<1x1xf32>, %arg2: vector<1x1xf32>) -> vector<1x1xf32> {
%d = nvgpu.mma.sync (%arg0, %arg1, %arg2) {mmaShape = [1, 64]} : (vector<2x1xf32>, vector<1x1xf32>, vector<1x1xf32>) -> vector<1x1xf32>
return %d : vector<1x1xf32>
}
```
commads:
mlir-opt test.mlir
stacktrace
```
mlir-opt:llvm-project/build/tools/mlir/include/mlir/Dialect/NVGPU/IR/NVGPUOps.h.inc:3612: std::array<int64_t, 3> mlir::nvgpu::MmaSyncOp::getMmaShapeAsArray(): Assertion `mmaShape.size() == 3 && "mmaShape should be three integers"' 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: ./bin/mlir-opt test.mlir
#0 0x0000564adbc54358 llvm::sys::PrintStackTrace(llvm::raw_ostream&, int) (./bin/mlir-opt+0x1e97358)
#1 0x0000564adbc51a15 llvm::sys::RunSignalHandlers() (./bin/mlir-opt+0x1e94a15)
#2 0x0000564adbc55396 SignalHandler(int, siginfo_t*, void*) Signals.cpp:0:0
#3 0x00007f124c900420 __restore_rt (/lib/x86_64-linux-gnu/libpthread.so.0+0x14420)
#4 0x00007f124bf5900b raise /build/glibc-B3wQXB/glibc-2.31/signal/../sysdeps/unix/sysv/linux/raise.c:51:1
#5 0x00007f124bf38859 abort /build/glibc-B3wQXB/glibc-2.31/stdlib/abort.c:81:7
#6 0x00007f124bf38729 get_sysdep_segment_value /build/glibc-B3wQXB/glibc-2.31/intl/loadmsgcat.c:509:8
#7 0x00007f124bf38729 _nl_load_domain /build/glibc-B3wQXB/glibc-2.31/intl/loadmsgcat.c:970:34
#8 0x00007f124bf49fd6 (/lib/x86_64-linux-gnu/libc.so.6+0x33fd6)
#9 0x0000564ade0fcb57 (./bin/mlir-opt+0x433fb57)
#10 0x0000564ae2fa2659 mlir::nvgpu::MmaSyncOp::verify() (./bin/mlir-opt+0x91e5659)
#11 0x0000564ae2ff12d3 mlir::Op<mlir::nvgpu::MmaSyncOp, mlir::OpTrait::ZeroRegions, mlir::OpTrait::OneResult, mlir::OpTrait::OneTypedResult<mlir::VectorType>::Impl, mlir::OpTrait::ZeroSuccessors, mlir::OpTrait::NOperands<3u>::Impl, mlir::OpTrait::OpInvariants, mlir::BytecodeOpInterface::Trait, mlir::ConditionallySpeculatable::Trait, mlir::OpTrait::AlwaysSpeculatableImplTrait, mlir::MemoryEffectOpInterface::Trait>::verifyInvariants(mlir::Operation*) (./bin/mlir-opt+0x92342d3)
#12 0x0000564ae2ff04ae mlir::RegisteredOperationName::Model<mlir::nvgpu::MmaSyncOp>::verifyInvariants(mlir::Operation*) (./bin/mlir-opt+0x92334ae)
#13 0x0000564ae4618a7f (anonymous namespace)::OperationVerifier::verifyOpAndDominance(mlir::Operation&) Verifier.cpp:0:0
#14 0x0000564ae4618c50 (anonymous namespace)::OperationVerifier::verifyOpAndDominance(mlir::Operation&) Verifier.cpp:0:0
#15 0x0000564ae4618146 mlir::verify(mlir::Operation*, bool) (./bin/mlir-opt+0xa85b146)
#16 0x0000564ae2cb8df5 mlir::parseAsmSourceFile(llvm::SourceMgr const&, mlir::Block*, mlir::ParserConfig const&, mlir::AsmParserState*, mlir::AsmParserCodeCompleteContext*) (./bin/mlir-opt+0x8efbdf5)
#17 0x0000564ae2c9d4d3 mlir::parseSourceFile(std::shared_ptr<llvm::SourceMgr> const&, mlir::Block*, mlir::ParserConfig const&, mlir::LocationAttr*) (./bin/mlir-opt+0x8ee04d3)
#18 0x0000564adbcf8afe mlir::parseSourceFileForTool(std::shared_ptr<llvm::SourceMgr> const&, mlir::ParserConfig const&, bool) (./bin/mlir-opt+0x1f3bafe)
#19 0x0000564adbcf7827 performActions(llvm::raw_ostream&, std::shared_ptr<llvm::SourceMgr> const&, mlir::MLIRContext*, mlir::MlirOptMainConfig const&) MlirOptMain.cpp:0:0
#20 0x0000564adbcf75a6 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&)::$_0>(long, std::unique_ptr<llvm::MemoryBuffer, std::default_delete<llvm::MemoryBuffer>>, llvm::MemoryBufferRef const&, llvm::raw_ostream&) MlirOptMain.cpp:0:0
#21 0x0000564ae4629855 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) (./bin/mlir-opt+0xa86c855)
#22 0x0000564adbced166 mlir::MlirOptMain(llvm::raw_ostream&, std::unique_ptr<llvm::MemoryBuffer, std::default_delete<llvm::MemoryBuffer>>, mlir::DialectRegistry&, mlir::MlirOptMainConfig const&) (./bin/mlir-opt+0x1f30166)
#23 0x0000564adbced4a2 mlir::MlirOptMain(int, char**, llvm::StringRef, llvm::StringRef, mlir::DialectRegistry&) (./bin/mlir-opt+0x1f304a2)
#24 0x0000564adbced805 mlir::MlirOptMain(int, char**, llvm::StringRef, mlir::DialectRegistry&) (./bin/mlir-opt+0x1f30805)
#25 0x0000564adbc3a9b3 main (./bin/mlir-opt+0x1e7d9b3)
#26 0x00007f124bf3a083 __libc_start_main /build/glibc-B3wQXB/glibc-2.31/csu/../csu/libc-start.c:342:3
#27 0x0000564adbc3a4ee _start (./bin/mlir-opt+0x1e7d4ee)
```
![]()
_______________________________________________
llvm-bugs mailing list
[email protected]
https://lists.llvm.org/cgi-bin/mailman/listinfo/llvm-bugs
