Skip to content

Support basic row major data movement + tilize/untilize #321 #323 #93

Support basic row major data movement + tilize/untilize #321 #323

Support basic row major data movement + tilize/untilize #321 #323 #93

GitHub Actions / MLIR Tests failed Aug 15, 2024 in 0s

43 tests run, 38 passed, 3 skipped, 2 failed.

Annotations

Check failure on line 1 in ttmlir/Dialect/TTMetal

See this annotation in the file changed.

@github-actions github-actions / MLIR Tests

ttmlir/Dialect/TTMetal.to_layout.mlir

Exit Code: 2
Raw output
Exit Code: 2

Command Output (stderr):
--
RUN: at line 1: /__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt --ttir-load-system-desc --ttir-implicit-device --ttir-allocate --convert-ttir-to-ttmetal /__w/tt-mlir/tt-mlir/test/ttmlir/Dialect/TTMetal/to_layout.mlir | /opt/ttmlir-toolchain/bin/FileCheck /__w/tt-mlir/tt-mlir/test/ttmlir/Dialect/TTMetal/to_layout.mlir
+ /opt/ttmlir-toolchain/bin/FileCheck /__w/tt-mlir/tt-mlir/test/ttmlir/Dialect/TTMetal/to_layout.mlir
+ /__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt --ttir-load-system-desc --ttir-implicit-device --ttir-allocate --convert-ttir-to-ttmetal /__w/tt-mlir/tt-mlir/test/ttmlir/Dialect/TTMetal/to_layout.mlir
ttmlir-opt: /__w/tt-mlir/tt-mlir/lib/Dialect/TTMetal/Transforms/Passes.cpp:94: mlir::tt::ttmetal::PhysicalCoordMapping::PhysicalCoordMapping(ArrayRef<tt::ChipDescAttr>): Assertion `chipPhysicalCores.getWorker().size() == static_cast<size_t>(grid[0] * grid[1])' failed.
PLEASE submit a bug report to https://github.com/llvm/llvm-project/issues/ and include the crash backtrace.
Stack dump:
0.	Program arguments: /__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt --ttir-load-system-desc --ttir-implicit-device --ttir-allocate --convert-ttir-to-ttmetal /__w/tt-mlir/tt-mlir/test/ttmlir/Dialect/TTMetal/to_layout.mlir
 #0 0x00007fd96af7cb08 llvm::sys::PrintStackTrace(llvm::raw_ostream&, int) (/opt/ttmlir-toolchain/lib/libLLVM.so.19.0git+0xf39b08)
 #1 0x00007fd96af7ad31 llvm::sys::RunSignalHandlers() (/opt/ttmlir-toolchain/lib/libLLVM.so.19.0git+0xf37d31)
 #2 0x00007fd96af7b8fd SignalHandler(int) Signals.cpp:0:0
 #3 0x00007fd969b1e520 (/lib/x86_64-linux-gnu/libc.so.6+0x42520)
 #4 0x00007fd969b729fc pthread_kill (/lib/x86_64-linux-gnu/libc.so.6+0x969fc)
 #5 0x00007fd969b1e476 gsignal (/lib/x86_64-linux-gnu/libc.so.6+0x42476)
 #6 0x00007fd969b047f3 abort (/lib/x86_64-linux-gnu/libc.so.6+0x287f3)
 #7 0x00007fd969b0471b (/lib/x86_64-linux-gnu/libc.so.6+0x2871b)
 #8 0x00007fd969b15e96 (/lib/x86_64-linux-gnu/libc.so.6+0x39e96)
 #9 0x0000557e25c54d83 mlir::tt::ttmetal::PhysicalCoordMapping::PhysicalCoordMapping(llvm::ArrayRef<mlir::tt::ChipDescAttr>) (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x2820d83)
#10 0x0000557e25c5312d mlir::tt::ttmetal::TTIRToTTMetalLayoutRewriter::relayout(mlir::tt::ttir::ToLayoutOp, mlir::PatternRewriter&) const (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x281f12d)
#11 0x0000557e25c52503 mlir::tt::ttmetal::TTIRToTTMetalLayoutRewriter::matchAndRewrite(mlir::tt::ttir::ToLayoutOp, mlir::PatternRewriter&) const (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x281e503)
#12 0x0000557e25a8e70a void llvm::function_ref<void ()>::callback_fn<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'()>(long) PatternApplicator.cpp:0:0
#13 0x0000557e25a8dccc 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&)>) (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x2659ccc)
#14 0x0000557e25a6dfe3 (anonymous namespace)::GreedyPatternRewriteDriver::processWorklist() GreedyPatternRewriteDriver.cpp:0:0
#15 0x0000557e25a6e111 void llvm::function_ref<void ()>::callback_fn<(anonymous namespace)::RegionPatternRewriteDriver::simplify(bool*) &&::'lambda2'()>(long) GreedyPatternRewriteDriver.cpp:0:0
#16 0x0000557e25a6ce77 mlir::applyPatternsAndFoldGreedily(mlir::Region&, mlir::FrozenRewritePatternSet const&, mlir::GreedyRewriteConfig, bool*) (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x2638e77)
#17 0x0000557e25c50efb mlir::tt::ttmetal::ConvertTTIRToTTMetal::runOnOperation() (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x281cefb)
#18 0x0000557e25b15396 void llvm::function_ref<void ()>::callback_fn<mlir::detail::OpToOpPassAdaptor::run(mlir::Pass*, mlir::Operation*, mlir::AnalysisManager, bool, unsigned int)::'lambda0'()>(long) Pass.cpp:0:0
#19 0x0000557e25b137b5 mlir::detail::OpToOpPassAdaptor::run(mlir::Pass*, mlir::Operation*, mlir::AnalysisManager, bool, unsigned int) (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x26df7b5)
#20 0x0000557e25b1393c mlir::detail::OpToOpPassAdaptor::runPipeline(mlir::OpPassManager&, mlir::Operation*, mlir::AnalysisManager, bool, unsigned int, mlir::PassInstrumentor*, mlir::PassInstrumentation::PipelineParentInfo const*) (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x26df93c)
#21 0x0000557e25b13a9d mlir::PassManager::runPasses(mlir::Operation*, mlir::AnalysisManager) (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x26dfa9d)
#22 0x0000557e25b13d73 mlir::PassManager::run(mlir::Operation*) (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x26dfd73)
#23 0x0000557e250962b5 performActions(llvm::raw_ostream&, std::shared_ptr<llvm::SourceMgr> const&, mlir::MLIRContext*, mlir::MlirOptMainConfig const&) MlirOptMain.cpp:0:0
#24 0x0000557e2509674d processBuffer(llvm::raw_ostream&, std::unique_ptr<llvm::MemoryBuffer, std::default_delete<llvm::MemoryBuffer>>, mlir::MlirOptMainConfig const&, mlir::DialectRegistry&, llvm::ThreadPoolInterface*) MlirOptMain.cpp:0:0
#25 0x0000557e25096829 llvm::LogicalResult llvm::function_ref<llvm::LogicalResult (std::unique_ptr<llvm::MemoryBuffer, std::default_delete<llvm::MemoryBuffer>>, 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::raw_ostream&)>(long, std::unique_ptr<llvm::MemoryBuffer, std::default_delete<llvm::MemoryBuffer>>, llvm::raw_ostream&) MlirOptMain.cpp:0:0
#26 0x0000557e25c4e920 llvm::function_ref<llvm::LogicalResult (std::unique_ptr<llvm::MemoryBuffer, std::default_delete<llvm::MemoryBuffer>>, llvm::raw_ostream&)>::operator()(std::unique_ptr<llvm::MemoryBuffer, std::default_delete<llvm::MemoryBuffer>>, llvm::raw_ostream&) const (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x281a920)
#27 0x0000557e25c4eb66 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::raw_ostream&)>, llvm::raw_ostream&, llvm::StringRef, llvm::StringRef) (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x281ab66)
#28 0x0000557e25092e5a mlir::MlirOptMain(llvm::raw_ostream&, std::unique_ptr<llvm::MemoryBuffer, std::default_delete<llvm::MemoryBuffer>>, mlir::DialectRegistry&, mlir::MlirOptMainConfig const&) (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x1c5ee5a)
#29 0x0000557e25096a3b mlir::MlirOptMain(int, char**, llvm::StringRef, llvm::StringRef, mlir::DialectRegistry&) (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x1c62a3b)
#30 0x0000557e25096bc1 mlir::MlirOptMain(int, char**, llvm::StringRef, mlir::DialectRegistry&) (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x1c62bc1)
#31 0x0000557e241a9225 main (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0xd75225)
#32 0x00007fd969b05d90 (/lib/x86_64-linux-gnu/libc.so.6+0x29d90)
#33 0x00007fd969b05e40 __libc_start_main (/lib/x86_64-linux-gnu/libc.so.6+0x29e40)
#34 0x0000557e241a9115 _start (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0xd75115)
FileCheck error: '<stdin>' is empty.
FileCheck command line:  /opt/ttmlir-toolchain/bin/FileCheck /__w/tt-mlir/tt-mlir/test/ttmlir/Dialect/TTMetal/to_layout.mlir

--

Check failure on line 1 in ttmlir/Silicon/TTMetal

See this annotation in the file changed.

@github-actions github-actions / MLIR Tests

ttmlir/Silicon/TTMetal.to_layout.mlir

Exit Code: 2
Raw output
Exit Code: 2

Command Output (stderr):
--
RUN: at line 1: /__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt --ttir-load-system-desc="path=" --ttir-implicit-device --ttir-allocate --convert-ttir-to-ttmetal --ttmetal-serialize-to-binary="output=/__w/tt-mlir/tt-mlir/build/test/ttmlir/Silicon/TTMetal/Output/to_layout.mlir.tmp.ttm" /__w/tt-mlir/tt-mlir/test/ttmlir/Silicon/TTMetal/to_layout.mlir | /opt/ttmlir-toolchain/bin/FileCheck /__w/tt-mlir/tt-mlir/test/ttmlir/Silicon/TTMetal/to_layout.mlir
+ /__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt --ttir-load-system-desc=path= --ttir-implicit-device --ttir-allocate --convert-ttir-to-ttmetal --ttmetal-serialize-to-binary=output=/__w/tt-mlir/tt-mlir/build/test/ttmlir/Silicon/TTMetal/Output/to_layout.mlir.tmp.ttm /__w/tt-mlir/tt-mlir/test/ttmlir/Silicon/TTMetal/to_layout.mlir
+ /opt/ttmlir-toolchain/bin/FileCheck /__w/tt-mlir/tt-mlir/test/ttmlir/Silicon/TTMetal/to_layout.mlir
ttmlir-opt: /__w/tt-mlir/tt-mlir/lib/Dialect/TTMetal/Transforms/Passes.cpp:94: mlir::tt::ttmetal::PhysicalCoordMapping::PhysicalCoordMapping(ArrayRef<tt::ChipDescAttr>): Assertion `chipPhysicalCores.getWorker().size() == static_cast<size_t>(grid[0] * grid[1])' failed.
PLEASE submit a bug report to https://github.com/llvm/llvm-project/issues/ and include the crash backtrace.
Stack dump:
0.	Program arguments: /__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt --ttir-load-system-desc=path= --ttir-implicit-device --ttir-allocate --convert-ttir-to-ttmetal --ttmetal-serialize-to-binary=output=/__w/tt-mlir/tt-mlir/build/test/ttmlir/Silicon/TTMetal/Output/to_layout.mlir.tmp.ttm /__w/tt-mlir/tt-mlir/test/ttmlir/Silicon/TTMetal/to_layout.mlir
 #0 0x00007f8f8f9f9b08 llvm::sys::PrintStackTrace(llvm::raw_ostream&, int) (/opt/ttmlir-toolchain/lib/libLLVM.so.19.0git+0xf39b08)
 #1 0x00007f8f8f9f7d31 llvm::sys::RunSignalHandlers() (/opt/ttmlir-toolchain/lib/libLLVM.so.19.0git+0xf37d31)
 #2 0x00007f8f8f9f88fd SignalHandler(int) Signals.cpp:0:0
 #3 0x00007f8f8e59b520 (/lib/x86_64-linux-gnu/libc.so.6+0x42520)
 #4 0x00007f8f8e5ef9fc pthread_kill (/lib/x86_64-linux-gnu/libc.so.6+0x969fc)
 #5 0x00007f8f8e59b476 gsignal (/lib/x86_64-linux-gnu/libc.so.6+0x42476)
 #6 0x00007f8f8e5817f3 abort (/lib/x86_64-linux-gnu/libc.so.6+0x287f3)
 #7 0x00007f8f8e58171b (/lib/x86_64-linux-gnu/libc.so.6+0x2871b)
 #8 0x00007f8f8e592e96 (/lib/x86_64-linux-gnu/libc.so.6+0x39e96)
 #9 0x00005644f4846d83 mlir::tt::ttmetal::PhysicalCoordMapping::PhysicalCoordMapping(llvm::ArrayRef<mlir::tt::ChipDescAttr>) (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x2820d83)
#10 0x00005644f484512d mlir::tt::ttmetal::TTIRToTTMetalLayoutRewriter::relayout(mlir::tt::ttir::ToLayoutOp, mlir::PatternRewriter&) const (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x281f12d)
#11 0x00005644f4844503 mlir::tt::ttmetal::TTIRToTTMetalLayoutRewriter::matchAndRewrite(mlir::tt::ttir::ToLayoutOp, mlir::PatternRewriter&) const (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x281e503)
#12 0x00005644f468070a void llvm::function_ref<void ()>::callback_fn<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'()>(long) PatternApplicator.cpp:0:0
#13 0x00005644f467fccc 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&)>) (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x2659ccc)
#14 0x00005644f465ffe3 (anonymous namespace)::GreedyPatternRewriteDriver::processWorklist() GreedyPatternRewriteDriver.cpp:0:0
#15 0x00005644f4660111 void llvm::function_ref<void ()>::callback_fn<(anonymous namespace)::RegionPatternRewriteDriver::simplify(bool*) &&::'lambda2'()>(long) GreedyPatternRewriteDriver.cpp:0:0
#16 0x00005644f465ee77 mlir::applyPatternsAndFoldGreedily(mlir::Region&, mlir::FrozenRewritePatternSet const&, mlir::GreedyRewriteConfig, bool*) (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x2638e77)
#17 0x00005644f4842efb mlir::tt::ttmetal::ConvertTTIRToTTMetal::runOnOperation() (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x281cefb)
#18 0x00005644f4707396 void llvm::function_ref<void ()>::callback_fn<mlir::detail::OpToOpPassAdaptor::run(mlir::Pass*, mlir::Operation*, mlir::AnalysisManager, bool, unsigned int)::'lambda0'()>(long) Pass.cpp:0:0
#19 0x00005644f47057b5 mlir::detail::OpToOpPassAdaptor::run(mlir::Pass*, mlir::Operation*, mlir::AnalysisManager, bool, unsigned int) (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x26df7b5)
#20 0x00005644f470593c mlir::detail::OpToOpPassAdaptor::runPipeline(mlir::OpPassManager&, mlir::Operation*, mlir::AnalysisManager, bool, unsigned int, mlir::PassInstrumentor*, mlir::PassInstrumentation::PipelineParentInfo const*) (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x26df93c)
#21 0x00005644f4705a9d mlir::PassManager::runPasses(mlir::Operation*, mlir::AnalysisManager) (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x26dfa9d)
#22 0x00005644f4705d73 mlir::PassManager::run(mlir::Operation*) (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x26dfd73)
#23 0x00005644f3c882b5 performActions(llvm::raw_ostream&, std::shared_ptr<llvm::SourceMgr> const&, mlir::MLIRContext*, mlir::MlirOptMainConfig const&) MlirOptMain.cpp:0:0
#24 0x00005644f3c8874d processBuffer(llvm::raw_ostream&, std::unique_ptr<llvm::MemoryBuffer, std::default_delete<llvm::MemoryBuffer>>, mlir::MlirOptMainConfig const&, mlir::DialectRegistry&, llvm::ThreadPoolInterface*) MlirOptMain.cpp:0:0
#25 0x00005644f3c88829 llvm::LogicalResult llvm::function_ref<llvm::LogicalResult (std::unique_ptr<llvm::MemoryBuffer, std::default_delete<llvm::MemoryBuffer>>, 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::raw_ostream&)>(long, std::unique_ptr<llvm::MemoryBuffer, std::default_delete<llvm::MemoryBuffer>>, llvm::raw_ostream&) MlirOptMain.cpp:0:0
#26 0x00005644f4840920 llvm::function_ref<llvm::LogicalResult (std::unique_ptr<llvm::MemoryBuffer, std::default_delete<llvm::MemoryBuffer>>, llvm::raw_ostream&)>::operator()(std::unique_ptr<llvm::MemoryBuffer, std::default_delete<llvm::MemoryBuffer>>, llvm::raw_ostream&) const (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x281a920)
#27 0x00005644f4840b66 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::raw_ostream&)>, llvm::raw_ostream&, llvm::StringRef, llvm::StringRef) (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x281ab66)
#28 0x00005644f3c84e5a mlir::MlirOptMain(llvm::raw_ostream&, std::unique_ptr<llvm::MemoryBuffer, std::default_delete<llvm::MemoryBuffer>>, mlir::DialectRegistry&, mlir::MlirOptMainConfig const&) (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x1c5ee5a)
#29 0x00005644f3c88a3b mlir::MlirOptMain(int, char**, llvm::StringRef, llvm::StringRef, mlir::DialectRegistry&) (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x1c62a3b)
#30 0x00005644f3c88bc1 mlir::MlirOptMain(int, char**, llvm::StringRef, mlir::DialectRegistry&) (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0x1c62bc1)
#31 0x00005644f2d9b225 main (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0xd75225)
#32 0x00007f8f8e582d90 (/lib/x86_64-linux-gnu/libc.so.6+0x29d90)
#33 0x00007f8f8e582e40 __libc_start_main (/lib/x86_64-linux-gnu/libc.so.6+0x29e40)
#34 0x00005644f2d9b115 _start (/__w/tt-mlir/tt-mlir/build/bin/ttmlir-opt+0xd75115)
FileCheck error: '<stdin>' is empty.
FileCheck command line:  /opt/ttmlir-toolchain/bin/FileCheck /__w/tt-mlir/tt-mlir/test/ttmlir/Silicon/TTMetal/to_layout.mlir

--