Closed ksriv001 closed 1 week ago
FYI @TT-billteng @tt-rkim
@ksriv001 some questions
PYTHONPATH
or TT_METAL_HOME
?v0.50.0
?v0.50.0
is quite old. We're planning to cut a release soon once models team can update their cadence. Is there any need to tie yourself to it? Can you use some of the prerelease/release candidates?@tt-rkim
I can try out the updated whl with one of the latest release candidates too and post my results here.
That is strange... can you reproduce your full command line output including installation into venv, clone etc? It is not finding the RISC-V binaries in the place I expect.
@tt-rkim Here is a fresh list of commands that I used to reproduce this error. There is one good difference that I am able to run the basic ops test successfully, but the mixtral test still gives me the same error. Here are the steps to reproduce it with the full error log:
Using whls from v51-rc11
TT_METAL_VERSION=v0.51.0-rc11
ARCH_NAME=wormhole_b0
pip install https://github.com/tenstorrent/tt-metal/releases/download/${TT_METAL_VERSION}/metal_libs-$(echo ${TT_METAL_VERSION} | sed -e 's/[v-]//g')+$(echo ${ARCH_NAME} | sed -e 's/_/./g')-cp38-cp38-linux_x86_64.whl
Cloning Repo git clone https://github.com/tenstorrent/tt-metal.git --recurse-submodules cd tt-metal git checkout dee91cf git submodule update --init --recursive git submodule foreach 'git lfs fetch --all && git lfs pull’
cd tt-metal
export PYTHONPATH=$(pwd) pip install -r tt_metal/python_env/requirements-dev.txt
Setup Mixtral export TT_METAL_ASYNC_DEVICE_QUEUE=1 export WH_ARCH_YAML=wormhole_b0_80_arch_eth_dispatch.yaml
export MIXTRAL_CKPT_DIR=/mnt/kanishk-mldata/mixtral/mixtral_consolidated/Mixtral-8x7B-v0.1/repacked_weights_dir export MIXTRAL_TOKENIZER_PATH=/mnt/kanishk-mldata/mixtral/mixtral_consolidated/Mixtral-8x7B-v0.1 export MIXTRAL_CACHE_PATH=/mnt/kanishk-mldata/mixtral/mixtral_consolidated/Mixtral-8x7B-v0.1/weights_cache_dir
Run Mixtral pytest -svv models/demos/t3000/mixtral8x7b/demo/demo.py::test_mixtral8x7b_demo[wormhole_b0-True-general_weights]
Error log:
user@llama31-70b-bade2922-deployment-5fc97999c7-vctst:~/tt-metal$ pytest -svv models/demos/t3000/mixtral8x7b/demo/demo.py::test_mixtral8x7b_demo[wormhole_b0-True-general_weights]
2024-08-22 20:51:57.637 | DEBUG | ttnn:<module>:83 - Initial ttnn.CONFIG:
Config{cache_path=/home/user/.cache/ttnn,model_cache_path=/home/user/.cache/ttnn/models,tmp_dir=/tmp/ttnn,enable_model_cache=false,enable_fast_runtime_mode=true,throw_exception_on_fallback=false,enable_logging=false,enable_graph_report=false,enable_detailed_buffer_report=false,enable_detailed_tensor_report=false,enable_comparison_mode=false,comparison_mode_pcc=0.9999,root_report_path=generated/ttnn/reports,report_name=std::nullopt,std::nullopt}
2024-08-22 20:51:57.708 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.polyval be migrated to C++?
2024-08-22 20:51:57.709 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.pearson_correlation_coefficient be migrated to C++?
2024-08-22 20:51:57.709 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.Conv1d be migrated to C++?
2024-08-22 20:51:57.713 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.conv2d be migrated to C++?
2024-08-22 20:51:57.713 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.reshape be migrated to C++?
2024-08-22 20:51:57.713 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.unsqueeze_to_4D be migrated to C++?
2024-08-22 20:51:57.713 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.squeeze be migrated to C++?
2024-08-22 20:51:57.714 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.from_torch be migrated to C++?
2024-08-22 20:51:57.714 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.to_torch be migrated to C++?
2024-08-22 20:51:57.714 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.to_device be migrated to C++?
2024-08-22 20:51:57.714 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.from_device be migrated to C++?
2024-08-22 20:51:57.714 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.allocate_tensor_on_device be migrated to C++?
2024-08-22 20:51:57.714 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.copy_host_to_device_tensor be migrated to C++?
2024-08-22 20:51:57.714 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.deallocate be migrated to C++?
2024-08-22 20:51:57.714 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.clone be migrated to C++?
2024-08-22 20:51:57.714 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.reallocate be migrated to C++?
2024-08-22 20:51:57.714 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.load_tensor be migrated to C++?
2024-08-22 20:51:57.714 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.dump_tensor be migrated to C++?
2024-08-22 20:51:57.714 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.as_tensor be migrated to C++?
2024-08-22 20:51:57.715 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.arange be migrated to C++?
2024-08-22 20:51:57.715 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.mse_loss be migrated to C++?
2024-08-22 20:51:57.715 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.l1_loss be migrated to C++?
2024-08-22 20:51:57.715 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.matmul be migrated to C++?
2024-08-22 20:51:57.715 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.linear be migrated to C++?
2024-08-22 20:51:57.716 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.max_pool2d be migrated to C++?
2024-08-22 20:51:57.716 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.max_pool2d_legacy be migrated to C++?
2024-08-22 20:51:57.716 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.global_avg_pool2d be migrated to C++?
2024-08-22 20:51:57.716 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.avg_pool2d be migrated to C++?
2024-08-22 20:51:57.716 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.mac be migrated to C++?
2024-08-22 20:51:57.717 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.lerp be migrated to C++?
2024-08-22 20:51:57.719 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.logit be migrated to C++?
2024-08-22 20:51:57.719 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.polygamma be migrated to C++?
2024-08-22 20:51:57.719 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.hardshrink be migrated to C++?
2024-08-22 20:51:57.719 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.celu be migrated to C++?
2024-08-22 20:51:57.719 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.softshrink be migrated to C++?
2024-08-22 20:51:57.720 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.clip be migrated to C++?
2024-08-22 20:51:57.720 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.threshold be migrated to C++?
2024-08-22 20:51:57.720 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.glu be migrated to C++?
2024-08-22 20:51:57.720 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.reglu be migrated to C++?
2024-08-22 20:51:57.720 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.swiglu be migrated to C++?
2024-08-22 20:51:57.720 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.geglu be migrated to C++?
2024-08-22 20:51:57.720 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.matmul be migrated to C++?
2024-08-22 20:51:57.721 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.linear be migrated to C++?
2024-08-22 20:51:57.721 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.conv2d be migrated to C++?
2024-08-22 20:51:57.721 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.max_pool2d be migrated to C++?
2024-08-22 20:51:57.721 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.max_pool2d_legacy be migrated to C++?
2024-08-22 20:51:57.721 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.global_avg_pool2d be migrated to C++?
2024-08-22 20:51:57.722 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.avg_pool2d be migrated to C++?
2024-08-22 20:51:57.722 | WARNING | ttnn.decorators:operation_decorator:790 - Should ttnn.Conv1d be migrated to C++?
============================================================================================================================= test session starts ==============================================================================================================================
platform linux -- Python 3.8.10, pytest-7.2.2, pluggy-1.5.0 -- /usr/bin/python3
cachedir: .pytest_cache
rootdir: /home/user/tt-metal, configfile: pytest.ini
plugins: dash-2.15.0, anyio-4.4.0, xdist-3.6.1, split-0.8.2, timeout-2.2.0
timeout: 300.0s
timeout method: signal
timeout func_only: False
collected 1 item
models/demos/t3000/mixtral8x7b/demo/demo.py::test_mixtral8x7b_demo[wormhole_b0-True-general_weights] Device | INFO | Opening user mode device driver
Detecting chips (found 8)
2024-08-22 20:51:58.829 | INFO | SiliconDriver - Detected 4 PCI devices : [4, 5, 6, 7]
2024-08-22 20:51:58.890 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 0)
2024-08-22 20:51:58.890 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 4)
2024-08-22 20:51:58.894 | INFO | SiliconDriver - Detected 4 PCI devices : [4, 5, 6, 7]
2024-08-22 20:51:58.903 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 1)
2024-08-22 20:51:58.904 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 5)
2024-08-22 20:51:58.907 | INFO | SiliconDriver - Detected 4 PCI devices : [4, 5, 6, 7]
2024-08-22 20:51:58.916 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 2)
2024-08-22 20:51:58.916 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 6)
2024-08-22 20:51:58.920 | INFO | SiliconDriver - Detected 4 PCI devices : [4, 5, 6, 7]
2024-08-22 20:51:58.929 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 3)
2024-08-22 20:51:58.930 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 7)
Metal | INFO | Initializing device 0. Program cache is NOT enabled
Metal | INFO | AI CLK for device 0 is: 1000 MHz
Metal | INFO | Initializing device 4. Program cache is NOT enabled
Metal | INFO | AI CLK for device 4 is: 1000 MHz
Metal | INFO | Initializing device 1. Program cache is NOT enabled
Metal | INFO | AI CLK for device 1 is: 1000 MHz
Metal | INFO | Initializing device 5. Program cache is NOT enabled
Metal | INFO | AI CLK for device 5 is: 1000 MHz
Metal | INFO | Initializing device 2. Program cache is NOT enabled
Metal | INFO | AI CLK for device 2 is: 1000 MHz
Metal | INFO | Initializing device 6. Program cache is NOT enabled
Metal | INFO | AI CLK for device 6 is: 1000 MHz
Metal | INFO | Initializing device 3. Program cache is NOT enabled
Metal | INFO | AI CLK for device 3 is: 1000 MHz
Metal | INFO | Initializing device 7. Program cache is NOT enabled
Metal | INFO | AI CLK for device 7 is: 1000 MHz
Metal | INFO | MMIO Device 0 : Tunnel 0 : Device 0
Metal | INFO | MMIO Device 0 : Tunnel 0 : Device 4
BuildKernels | INFO | erisc compile failure -- cmd: cd /usr/local/lib/python3.8/dist-packages/built/6272/kernels/eth_tunneler/11703499288861771092/erisc/ && /usr/local/lib/python3.8/dist-packages/tt_metal/third_party/sfpi/compiler/bin/riscv32-unknown-elf-g++ -mwormhole -march=rv32imw -mtune=rvtt-b1 -mabi=ilp32 -std=c++17 -flto -ffast-math -fno-use-cxa-atexit -fno-exceptions -Wall -Werror -Wno-unknown-pragmas -Wno-error=multistatement-macros -Wno-error=parentheses -Wno-error=unused-but-set-variable -Wno-unused-variable -Wno-unused-function -Os -fno-delete-null-pointer-checks -DARCH_WORMHOLE -DTENSIX_FIRMWARE -DLOCAL_MEM_EN=0 -DCOMPILE_FOR_ERISC -DERISC -DRISC_B0_HW -DKERNEL_BUILD -DDISPATCH_KERNEL=1 -DDOWNSTREAM_NOC_X=0 -DDOWNSTREAM_NOC_Y=0 -DMY_NOC_X=0 -DMY_NOC_Y=0 -DSKIP_NOC_LOGGING=1 -DUPSTREAM_NOC_X=0 -DUPSTREAM_NOC_Y=0 -DNOC_INDEX=0 -DKERNEL_COMPILE_TIME_ARG_0=3670727370 -DKERNEL_COMPILE_TIME_ARG_1=2 -DKERNEL_COMPILE_TIME_ARG_2=6400 -DKERNEL_COMPILE_TIME_ARG_3=4096 -DKERNEL_COMPILE_TIME_ARG_4=65539 -DKERNEL_COMPILE_TIME_ARG_5=33621505 -DKERNEL_COMPILE_TIME_ARG_6=6742 -DKERNEL_COMPILE_TIME_ARG_7=4096 -DKERNEL_COMPILE_TIME_ARG_8=10496 -DKERNEL_COMPILE_TIME_ARG_9=4096 -DKERNEL_COMPILE_TIME_ARG_10=33687041 -DKERNEL_COMPILE_TIME_ARG_11=65543 -DKERNEL_COMPILE_TIME_ARG_12=233472 -DKERNEL_COMPILE_TIME_ARG_13=28672 -DKERNEL_COMPILE_TIME_ARG_14=0 -DKERNEL_COMPILE_TIME_ARG_15=0 -I. -I.. -I/usr/local/lib/python3.8/dist-packages/ -I/usr/local/lib/python3.8/dist-packages/tt_metal -I/usr/local/lib/python3.8/dist-packages/tt_metal/include -I/usr/local/lib/python3.8/dist-packages/tt_metal/hw/inc -I/usr/local/lib/python3.8/dist-packages/tt_metal/hw/inc/debug -I/usr/local/lib/python3.8/dist-packages/tt_metal/hw/inc/wormhole -I/usr/local/lib/python3.8/dist-packages/tt_metal/hw/inc/wormhole/wormhole_b0_defines -I/usr/local/lib/python3.8/dist-packages/tt_metal/hw/inc/wormhole/noc -I/usr/local/lib/python3.8/dist-packages/tt_metal/third_party/umd/device/wormhole_b0 -I/usr/local/lib/python3.8/dist-packages/tt_metal/hw/ckernels/wormhole_b0/metal/common -I/usr/local/lib/python3.8/dist-packages/tt_metal/hw/ckernels/wormhole_b0/metal/llk_io -I/usr/local/lib/python3.8/dist-packages/tt_metal/third_party/tt_llk_wormhole_b0/common/inc -I/usr/local/lib/python3.8/dist-packages/tt_metal/third_party/tt_llk_wormhole_b0/llk_lib -I /usr/local/lib/python3.8/dist-packages/tt_metal/hw/ckernels/wormhole_b0/metal/common -I /usr/local/lib/python3.8/dist-packages/tt_metal/hw/ckernels/wormhole_b0/metal/llk_io -I /usr/local/lib/python3.8/dist-packages/tt_metal/hw/inc/ethernet -c -o erisck.o /usr/local/lib/python3.8/dist-packages/tt_metal/hw/firmware/src/erisck.cc
In file included from ../kernel_includes.hpp:1,
from /usr/local/lib/python3.8/dist-packages/tt_metal/hw/firmware/src/erisck.cc:23:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/eth_tunneler.cpp:10:10: fatal error: tests/tt_metal/tt_metal/perf_microbenchmark/routing/kernels/traffic_gen.hpp: No such file or directory
10 | #include "tests/tt_metal/tt_metal/perf_microbenchmark/routing/kernels/traffic_gen.hpp"
| ^~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~
compilation terminated.
Always | FATAL | erisc build failed
Always | FATAL | Failed to generate binaries for eth_tunneler TT_THROW @ /home/runner/work/tt-metal/tt-metal/tt_metal/jit_build/build.cpp:397: tt::exception
info:
erisc build failed
backtrace:
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xd199e) [0x7f669ed2699e]
--- tt::tt_metal::JitBuildState::compile_one(std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, tt::tt_metal::JitBuildSettings const*, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&) const
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xd7b43) [0x7f669ed2cb43]
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xdbe0a) [0x7f669ed30e0a]
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xda9ad) [0x7f669ed2f9ad]
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xd9eb9) [0x7f669ed2eeb9]
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xd9d1a) [0x7f669ed2ed1a]
--- /lib/x86_64-linux-gnu/libpthread.so.0(+0x8609) [0x7f675f7f9609]
--- /lib/x86_64-linux-gnu/libc.so.6(clone+0x43) [0x7f675f933353]
BuildKernels | INFO | idle_erisc compile failure -- cmd: cd /usr/local/lib/python3.8/dist-packages/built/6272/kernels/packet_demux/15166713051469033642/idle_erisc/ && /usr/local/lib/python3.8/dist-packages/tt_metal/third_party/sfpi/compiler/bin/riscv32-unknown-elf-g++ -mwormhole -march=rv32imw -mtune=rvtt-b1 -mabi=ilp32 -std=c++17 -flto -ffast-math -fno-use-cxa-atexit -fno-exceptions -Wall -Werror -Wno-unknown-pragmas -Wno-error=multistatement-macros -Wno-error=parentheses -Wno-error=unused-but-set-variable -Wno-unused-variable -Wno-unused-function -Os -fno-tree-loop-distribute-patterns -DARCH_WORMHOLE -DTENSIX_FIRMWARE -DLOCAL_MEM_EN=0 -DCOMPILE_FOR_IDLE_ERISC -DERISC -DRISC_B0_HW -DKERNEL_BUILD -DDISPATCH_KERNEL=1 -DDOWNSTREAM_NOC_X=0 -DDOWNSTREAM_NOC_Y=0 -DMY_NOC_X=0 -DMY_NOC_Y=0 -DSKIP_NOC_LOGGING=1 -DUPSTREAM_NOC_X=0 -DUPSTREAM_NOC_Y=0 -DNOC_INDEX=0 -DKERNEL_COMPILE_TIME_ARG_0=177 -DKERNEL_COMPILE_TIME_ARG_1=6742 -DKERNEL_COMPILE_TIME_ARG_2=4096 -DKERNEL_COMPILE_TIME_ARG_3=1 -DKERNEL_COMPILE_TIME_ARG_4=1544 -DKERNEL_COMPILE_TIME_ARG_5=0 -DKERNEL_COMPILE_TIME_ARG_6=0 -DKERNEL_COMPILE_TIME_ARG_7=0 -DKERNEL_COMPILE_TIME_ARG_8=6912 -DKERNEL_COMPILE_TIME_ARG_9=8192 -DKERNEL_COMPILE_TIME_ARG_10=0 -DKERNEL_COMPILE_TIME_ARG_11=0 -DKERNEL_COMPILE_TIME_ARG_12=0 -DKERNEL_COMPILE_TIME_ARG_13=0 -DKERNEL_COMPILE_TIME_ARG_14=0 -DKERNEL_COMPILE_TIME_ARG_15=0 -DKERNEL_COMPILE_TIME_ARG_16=1 -DKERNEL_COMPILE_TIME_ARG_17=0 -DKERNEL_COMPILE_TIME_ARG_18=2 -DKERNEL_COMPILE_TIME_ARG_19=0 -DKERNEL_COMPILE_TIME_ARG_20=0 -DKERNEL_COMPILE_TIME_ARG_21=84 -DKERNEL_COMPILE_TIME_ARG_22=0 -DKERNEL_COMPILE_TIME_ARG_23=0 -DKERNEL_COMPILE_TIME_ARG_24=0 -DKERNEL_COMPILE_TIME_ARG_25=1 -DKERNEL_COMPILE_TIME_ARG_26=268 -DKERNEL_COMPILE_TIME_ARG_27=0 -DKERNEL_COMPILE_TIME_ARG_28=0 -DKERNEL_COMPILE_TIME_ARG_29=0 -I. -I.. -I/usr/local/lib/python3.8/dist-packages/ -I/usr/local/lib/python3.8/dist-packages/tt_metal -I/usr/local/lib/python3.8/dist-packages/tt_metal/include -I/usr/local/lib/python3.8/dist-packages/tt_metal/hw/inc -I/usr/local/lib/python3.8/dist-packages/tt_metal/hw/inc/debug -I/usr/local/lib/python3.8/dist-packages/tt_metal/hw/inc/wormhole -I/usr/local/lib/python3.8/dist-packages/tt_metal/hw/inc/wormhole/wormhole_b0_defines -I/usr/local/lib/python3.8/dist-packages/tt_metal/hw/inc/wormhole/noc -I/usr/local/lib/python3.8/dist-packages/tt_metal/third_party/umd/device/wormhole_b0 -I/usr/local/lib/python3.8/dist-packages/tt_metal/hw/ckernels/wormhole_b0/metal/common -I/usr/local/lib/python3.8/dist-packages/tt_metal/hw/ckernels/wormhole_b0/metal/llk_io -I/usr/local/lib/python3.8/dist-packages/tt_metal/third_party/tt_llk_wormhole_b0/common/inc -I/usr/local/lib/python3.8/dist-packages/tt_metal/third_party/tt_llk_wormhole_b0/llk_lib -I /usr/local/lib/python3.8/dist-packages/tt_metal/hw/ckernels/wormhole_b0/metal/common -I /usr/local/lib/python3.8/dist-packages/tt_metal/hw/ckernels/wormhole_b0/metal/llk_io -I /usr/local/lib/python3.8/dist-packages/tt_metal/hw/firmware/src -c -o idle_erisck.o /usr/local/lib/python3.8/dist-packages/tt_metal/hw/firmware/src/idle_erisck.cc
In file included from ../kernel_includes.hpp:1,
from /usr/local/lib/python3.8/dist-packages/tt_metal/hw/firmware/src/idle_erisck.cc:23:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/packet_demux.cpp:9:10: fatal error: tests/tt_metal/tt_metal/perf_microbenchmark/routing/kernels/traffic_gen.hpp: No such file or directory
9 | #include "tests/tt_metal/tt_metal/perf_microbenchmark/routing/kernels/traffic_gen.hpp"
| ^~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~
compilation terminated.
Always | FATAL | idle_erisc build failed
Always | FATAL | Failed to generate binaries for packet_demux TT_THROW @ /home/runner/work/tt-metal/tt-metal/tt_metal/jit_build/build.cpp:397: tt::exception
info:
idle_erisc build failed
backtrace:
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xd199e) [0x7f669ed2699e]
--- tt::tt_metal::JitBuildState::compile_one(std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, tt::tt_metal::JitBuildSettings const*, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&) const
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xd7b43) [0x7f669ed2cb43]
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xdbe0a) [0x7f669ed30e0a]
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xda9ad) [0x7f669ed2f9ad]
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xd9eb9) [0x7f669ed2eeb9]
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xd9d1a) [0x7f669ed2ed1a]
--- /lib/x86_64-linux-gnu/libpthread.so.0(+0x8609) [0x7f675f7f9609]
--- /lib/x86_64-linux-gnu/libc.so.6(clone+0x43) [0x7f675f933353]
libc++abi: terminating due to uncaught exception of type std::runtime_error: TT_THROW @ /home/runner/work/tt-metal/tt-metal/tt_metal/impl/program/program.cpp:35: tt::exception
info:
Failed to generate binaries for packet_demux TT_THROW @ /home/runner/work/tt-metal/tt-metal/tt_metal/jit_build/build.cpp:397: tt::exception
info:
idle_erisc build failed
backtrace:
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xd199e) [0x7f669ed2699e]
--- tt::tt_metal::JitBuildState::compile_one(std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, tt::tt_metal::JitBuildSettings const*, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&) const
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xd7b43) [0x7f669ed2cb43]
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xdbe0a) [0x7f669ed30e0a]
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xda9ad) [0x7f669ed2f9ad]
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xd9eb9) [0x7f669ed2eeb9]
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xd9d1a) [0x7f669ed2ed1a]
--- /lib/x86_64-linux-gnu/libpthread.so.0(+0x8609) [0x7f675f7f9609]
--- /lib/x86_64-linux-gnu/libc.so.6(clone+0x43) [0x7f675f933353]
backtrace:
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0x19d94e) [0x7f669edf294e]
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xd7b43) [0x7f669ed2cb43]
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xdbe0a) [0x7f669ed30e0a]
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xda9ad) [0x7f669ed2f9ad]
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xd9eb9) [0x7f669ed2eeb9]
--- /usr/local/lib/python3.8/dist-packages/ttnn/build/lib/libtt_metal.so(+0xd9d1a) [0x7f669ed2ed1a]
--- /lib/x86_64-linux-gnu/libpthread.so.0(+0x8609) [0x7f675f7f9609]
--- /lib/x86_64-linux-gnu/libc.so.6(clone+0x43) [0x7f675f933353]
Fatal Python error: Aborted
Current thread 0x00007f675f648740 (most recent call first):
File "/usr/local/lib/python3.8/dist-packages/ttnn/multi_device.py", line 49 in open_device_mesh
File "/home/user/tt-metal/conftest.py", line 243 in t3k_device_mesh
File "/usr/local/lib/python3.8/dist-packages/_pytest/fixtures.py", line 901 in call_fixture_func
File "/usr/local/lib/python3.8/dist-packages/_pytest/fixtures.py", line 1129 in pytest_fixture_setup
File "/usr/local/lib/python3.8/dist-packages/pluggy/_callers.py", line 103 in _multicall
File "/usr/local/lib/python3.8/dist-packages/pluggy/_manager.py", line 120 in _hookexec
File "/usr/local/lib/python3.8/dist-packages/pluggy/_hooks.py", line 513 in __call__
File "/usr/local/lib/python3.8/dist-packages/_pytest/fixtures.py", line 1075 in execute
File "/usr/local/lib/python3.8/dist-packages/_pytest/fixtures.py", line 677 in _compute_fixture_value
File "/usr/local/lib/python3.8/dist-packages/_pytest/fixtures.py", line 591 in _get_active_fixturedef
File "/usr/local/lib/python3.8/dist-packages/_pytest/fixtures.py", line 569 in getfixturevalue
File "/usr/local/lib/python3.8/dist-packages/_pytest/fixtures.py", line 550 in _fillfixtures
File "/usr/local/lib/python3.8/dist-packages/_pytest/python.py", line 1792 in setup
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 492 in setup
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 155 in pytest_runtest_setup
File "/usr/local/lib/python3.8/dist-packages/pluggy/_callers.py", line 103 in _multicall
File "/usr/local/lib/python3.8/dist-packages/pluggy/_manager.py", line 120 in _hookexec
File "/usr/local/lib/python3.8/dist-packages/pluggy/_hooks.py", line 513 in __call__
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 260 in <lambda>
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 339 in from_call
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 259 in call_runtest_hook
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 220 in call_and_report
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 125 in runtestprotocol
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 112 in pytest_runtest_protocol
File "/usr/local/lib/python3.8/dist-packages/pluggy/_callers.py", line 103 in _multicall
File "/usr/local/lib/python3.8/dist-packages/pluggy/_manager.py", line 120 in _hookexec
File "/usr/local/lib/python3.8/dist-packages/pluggy/_hooks.py", line 513 in __call__
File "/usr/local/lib/python3.8/dist-packages/_pytest/main.py", line 349 in pytest_runtestloop
File "/usr/local/lib/python3.8/dist-packages/pluggy/_callers.py", line 103 in _multicall
File "/usr/local/lib/python3.8/dist-packages/pluggy/_manager.py", line 120 in _hookexec
File "/usr/local/lib/python3.8/dist-packages/pluggy/_hooks.py", line 513 in __call__
File "/usr/local/lib/python3.8/dist-packages/_pytest/main.py", line 324 in _main
File "/usr/local/lib/python3.8/dist-packages/_pytest/main.py", line 270 in wrap_session
File "/usr/local/lib/python3.8/dist-packages/_pytest/main.py", line 317 in pytest_cmdline_main
File "/usr/local/lib/python3.8/dist-packages/pluggy/_callers.py", line 103 in _multicall
File "/usr/local/lib/python3.8/dist-packages/pluggy/_manager.py", line 120 in _hookexec
File "/usr/local/lib/python3.8/dist-packages/pluggy/_hooks.py", line 513 in __call__
File "/usr/local/lib/python3.8/dist-packages/_pytest/config/__init__.py", line 167 in main
File "/usr/local/lib/python3.8/dist-packages/_pytest/config/__init__.py", line 190 in console_main
File "/usr/local/bin/pytest", line 8 in <module>
Aborted (core dumped)
@mtairum @aliuTT @tt-aho
Do any of you know why we're importing kernels from tests/
in mixtral7b?
@mtairum @aliuTT @tt-aho Do any of you know why we're importing kernels from
tests/
in mixtral7b?
This is a FD kernel, not Mixtral specific. I will take a look at removing it, I don't think it should be there
I've removed the import on latest main
@ksriv001 Once this completes and uploads the eager package artifacts, please find the wheels there and try it out: https://github.com/tenstorrent/tt-metal/actions/runs/10561048601
We will be waiting for your response.
Tried with the new whl, got these errors:
user@llama31-70b-4837c6bc-deployment-6d55668c84-b42xd:~/tt-metal$ pytest -svv models/demos/t3000/mixtral8x7b/demo/demo.py::test_mixtral8x7b_demo[wormhole_b0-True-general_weights]
2024-08-26 23:07:39.876 | DEBUG | ttnn.library_tweaks:prepare_dir_as_metal_home:54 - Existing installation of 0.51.0rc36.dev16+wormhole.b0 detected
2024-08-26 23:07:39.973 | DEBUG | ttnn:<module>:82 - Initial ttnn.CONFIG:
Config{cache_path=/home/user/.cache/ttnn,model_cache_path=/home/user/.cache/ttnn/models,tmp_dir=/tmp/ttnn,enable_model_cache=false,enable_fast_runtime_mode=true,throw_exception_on_fallback=false,enable_logging=false,enable_graph_report=false,enable_detailed_buffer_report=false,enable_detailed_tensor_report=false,enable_comparison_mode=false,comparison_mode_pcc=0.9999,root_report_path=generated/ttnn/reports,report_name=std::nullopt,std::nullopt}
2024-08-26 23:07:40.043 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.pearson_correlation_coefficient be migrated to C++?
2024-08-26 23:07:40.044 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.Conv1d be migrated to C++?
2024-08-26 23:07:40.047 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.conv2d be migrated to C++?
2024-08-26 23:07:40.047 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.reshape be migrated to C++?
2024-08-26 23:07:40.047 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.unsqueeze_to_4D be migrated to C++?
2024-08-26 23:07:40.047 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.squeeze be migrated to C++?
2024-08-26 23:07:40.047 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.from_torch be migrated to C++?
2024-08-26 23:07:40.047 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.to_torch be migrated to C++?
2024-08-26 23:07:40.048 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.to_device be migrated to C++?
2024-08-26 23:07:40.048 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.from_device be migrated to C++?
2024-08-26 23:07:40.048 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.allocate_tensor_on_device be migrated to C++?
2024-08-26 23:07:40.048 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.copy_host_to_device_tensor be migrated to C++?
2024-08-26 23:07:40.048 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.deallocate be migrated to C++?
2024-08-26 23:07:40.048 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.reallocate be migrated to C++?
2024-08-26 23:07:40.048 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.load_tensor be migrated to C++?
2024-08-26 23:07:40.048 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.dump_tensor be migrated to C++?
2024-08-26 23:07:40.048 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.as_tensor be migrated to C++?
2024-08-26 23:07:40.049 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.max_pool2d be migrated to C++?
2024-08-26 23:07:40.050 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.max_pool2d_legacy be migrated to C++?
2024-08-26 23:07:40.050 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.global_avg_pool2d be migrated to C++?
2024-08-26 23:07:40.050 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.avg_pool2d be migrated to C++?
2024-08-26 23:07:40.052 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.conv2d be migrated to C++?
2024-08-26 23:07:40.053 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.max_pool2d be migrated to C++?
2024-08-26 23:07:40.053 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.max_pool2d_legacy be migrated to C++?
2024-08-26 23:07:40.053 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.global_avg_pool2d be migrated to C++?
2024-08-26 23:07:40.053 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.avg_pool2d be migrated to C++?
2024-08-26 23:07:40.053 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.Conv1d be migrated to C++?
========================================================= test session starts =========================================================
platform linux -- Python 3.8.10, pytest-7.2.2, pluggy-1.5.0 -- /usr/bin/python3
cachedir: .pytest_cache
rootdir: /home/user/tt-metal, configfile: pytest.ini
plugins: anyio-4.4.0, dash-2.15.0, split-0.8.2, xdist-3.6.1, timeout-2.2.0
timeout: 300.0s
timeout method: signal
timeout func_only: False
collected 1 item
models/demos/t3000/mixtral8x7b/demo/demo.py::test_mixtral8x7b_demo[wormhole_b0-True-general_weights] Device | INFO | Opening user mode device driver
Detecting chips (found 8)
2024-08-26 23:07:41.395 | INFO | SiliconDriver - Detected 4 PCI devices : [0, 1, 2, 3]
2024-08-26 23:07:41.477 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 0)
2024-08-26 23:07:41.478 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 4)
2024-08-26 23:07:41.481 | INFO | SiliconDriver - Detected 4 PCI devices : [0, 1, 2, 3]
2024-08-26 23:07:41.496 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 1)
2024-08-26 23:07:41.496 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 5)
2024-08-26 23:07:41.500 | INFO | SiliconDriver - Detected 4 PCI devices : [0, 1, 2, 3]
2024-08-26 23:07:41.514 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 2)
2024-08-26 23:07:41.515 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 6)
2024-08-26 23:07:41.517 | INFO | SiliconDriver - Detected 4 PCI devices : [0, 1, 2, 3]
2024-08-26 23:07:41.532 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 3)
2024-08-26 23:07:41.533 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 7)
Metal | INFO | Initializing device 0. Program cache is NOT enabled
Metal | INFO | AI CLK for device 0 is: 1000 MHz
Metal | INFO | Initializing device 4. Program cache is NOT enabled
Metal | INFO | AI CLK for device 4 is: 1000 MHz
Metal | INFO | Initializing device 1. Program cache is NOT enabled
Metal | INFO | AI CLK for device 1 is: 1000 MHz
Metal | INFO | Initializing device 5. Program cache is NOT enabled
Metal | INFO | AI CLK for device 5 is: 1000 MHz
Metal | INFO | Initializing device 2. Program cache is NOT enabled
Metal | INFO | AI CLK for device 2 is: 1000 MHz
Metal | INFO | Initializing device 6. Program cache is NOT enabled
Metal | INFO | AI CLK for device 6 is: 1000 MHz
Metal | INFO | Initializing device 3. Program cache is NOT enabled
Metal | INFO | AI CLK for device 3 is: 1000 MHz
Metal | INFO | Initializing device 7. Program cache is NOT enabled
Metal | INFO | AI CLK for device 7 is: 1000 MHz
BuildKernels | INFO | ncrisc compile failure -- cmd: cd /home/user/tt-metal/.ttnn_runtime_artifacts/built/6160/kernels/cq_prefetch/14525334194422425254/ncrisc/ && /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/third_party/sfpi/compiler/bin/riscv32-unknown-elf-g++ -mwormhole -march=rv32imw -mtune=rvtt-b1 -mabi=ilp32 -std=c++17 -flto -ffast-math -fno-use-cxa-atexit -fno-exceptions -Wall -Werror -Wno-unknown-pragmas -Wno-error=multistatement-macros -Wno-error=parentheses -Wno-error=unused-but-set-variable -Wno-unused-variable -Wno-unused-function -Os -fno-tree-loop-distribute-patterns -DARCH_WORMHOLE -DTENSIX_FIRMWARE -DLOCAL_MEM_EN=0 -DCOMPILE_FOR_NCRISC -DKERNEL_BUILD -DDISPATCH_KERNEL=1 -DDOWNSTREAM_NOC_X=2 -DDOWNSTREAM_NOC_Y=10 -DFD_CORE_TYPE=0 -DMY_NOC_X=1 -DMY_NOC_Y=10 -DUPSTREAM_NOC_INDEX=0 -DUPSTREAM_NOC_X=0 -DUPSTREAM_NOC_Y=0 -DNOC_INDEX=0 -DKERNEL_COMPILE_TIME_ARG_0=110592 -DKERNEL_COMPILE_TIME_ARG_1=12 -DKERNEL_COMPILE_TIME_ARG_2=128 -DKERNEL_COMPILE_TIME_ARG_3=1 -DKERNEL_COMPILE_TIME_ARG_4=0 -DKERNEL_COMPILE_TIME_ARG_5=128 -DKERNEL_COMPILE_TIME_ARG_6=805310336 -DKERNEL_COMPILE_TIME_ARG_7=107520 -DKERNEL_COMPILE_TIME_ARG_8=3068 -DKERNEL_COMPILE_TIME_ARG_9=107360 -DKERNEL_COMPILE_TIME_ARG_10=107364 -DKERNEL_COMPILE_TIME_ARG_11=110592 -DKERNEL_COMPILE_TIME_ARG_12=262144 -DKERNEL_COMPILE_TIME_ARG_13=372736 -DKERNEL_COMPILE_TIME_ARG_14=131072 -DKERNEL_COMPILE_TIME_ARG_15=0 -DKERNEL_COMPILE_TIME_ARG_16=64 -DKERNEL_COMPILE_TIME_ARG_17=0 -DKERNEL_COMPILE_TIME_ARG_18=0 -DKERNEL_COMPILE_TIME_ARG_19=12 -DKERNEL_COMPILE_TIME_ARG_20=4 -DKERNEL_COMPILE_TIME_ARG_21=1 -DKERNEL_COMPILE_TIME_ARG_22=1 -I. -I.. -I/home/user/tt-metal/.ttnn_runtime_artifacts/ -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/include -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/inc -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/inc/debug -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/inc/wormhole -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/inc/wormhole/wormhole_b0_defines -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/inc/wormhole/noc -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/third_party/umd/device/wormhole_b0 -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/ckernels/wormhole_b0/metal/common -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/ckernels/wormhole_b0/metal/llk_io -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/third_party/tt_llk_wormhole_b0/common/inc -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/third_party/tt_llk_wormhole_b0/llk_lib -I /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src -I /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/ckernels/wormhole_b0/metal/common -I /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/ckernels/wormhole_b0/metal/llk_io -c -o ncrisck.o /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp: In function 'uint32_t process_relay_inline_cmd(uint32_t, uint32_t&)':
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:352:69: error: no matching function for call to 'cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(uint32_t&)'
352 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(npages);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:14,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id> void cb_release_pages(uint32_t)'
293 | void cb_release_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:352:69: error: narrowing conversion of '10272' from 'uint32_t' {aka 'long unsigned int'} to 'unsigned char' [-Wnarrowing]
352 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(npages);
| ^
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp: In function 'uint32_t process_relay_paged_cmd_large(uint32_t, uint32_t&, uint32_t, uint32_t, uint32_t, uint32_t, uint32_t)':
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:500:73: error: no matching function for call to 'cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(uint32_t&)'
500 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(npages);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:14,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id> void cb_release_pages(uint32_t)'
293 | void cb_release_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:500:73: error: narrowing conversion of '10272' from 'uint32_t' {aka 'long unsigned int'} to 'unsigned char' [-Wnarrowing]
500 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(npages);
| ^
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:516:77: error: no matching function for call to 'cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(long unsigned int)'
516 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(npages + 1);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:14,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id> void cb_release_pages(uint32_t)'
293 | void cb_release_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:516:77: error: narrowing conversion of '10272' from 'uint32_t' {aka 'long unsigned int'} to 'unsigned char' [-Wnarrowing]
516 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(npages + 1);
| ^
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:518:68: error: no matching function for call to 'cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(int)'
518 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(1);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:14,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id> void cb_release_pages(uint32_t)'
293 | void cb_release_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:518:68: error: narrowing conversion of '10272' from 'uint32_t' {aka 'long unsigned int'} to 'unsigned char' [-Wnarrowing]
518 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(1);
| ^
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp: In function 'uint32_t process_relay_paged_cmd(uint32_t, uint32_t&, uint32_t)':
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:616:73: error: no matching function for call to 'cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(uint32_t&)'
616 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(npages);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:14,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id> void cb_release_pages(uint32_t)'
293 | void cb_release_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:616:73: error: narrowing conversion of '10272' from 'uint32_t' {aka 'long unsigned int'} to 'unsigned char' [-Wnarrowing]
616 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(npages);
| ^
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:636:73: error: no matching function for call to 'cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(long unsigned int)'
636 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(npages + 1);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:14,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id> void cb_release_pages(uint32_t)'
293 | void cb_release_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:636:73: error: narrowing conversion of '10272' from 'uint32_t' {aka 'long unsigned int'} to 'unsigned char' [-Wnarrowing]
636 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(npages + 1);
| ^
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp: In function 'void process_relay_paged_packed_sub_cmds(uint32_t)':
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:735:73: error: no matching function for call to 'cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(uint32_t&)'
735 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(npages);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:14,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id> void cb_release_pages(uint32_t)'
293 | void cb_release_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:735:73: error: narrowing conversion of '10272' from 'uint32_t' {aka 'long unsigned int'} to 'unsigned char' [-Wnarrowing]
735 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(npages);
| ^
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:752:73: error: no matching function for call to 'cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(long unsigned int)'
752 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(npages + 1);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:14,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id> void cb_release_pages(uint32_t)'
293 | void cb_release_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:752:73: error: narrowing conversion of '10272' from 'uint32_t' {aka 'long unsigned int'} to 'unsigned char' [-Wnarrowing]
752 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(npages + 1);
| ^
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp: In function 'uint32_t process_relay_linear_cmd(uint32_t, uint32_t&)':
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:827:73: error: no matching function for call to 'cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(uint32_t&)'
827 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(npages);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:14,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id> void cb_release_pages(uint32_t)'
293 | void cb_release_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:827:73: error: narrowing conversion of '10272' from 'uint32_t' {aka 'long unsigned int'} to 'unsigned char' [-Wnarrowing]
827 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(npages);
| ^
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:844:73: error: no matching function for call to 'cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(long unsigned int)'
844 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(npages + 1);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:14,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id> void cb_release_pages(uint32_t)'
293 | void cb_release_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:844:73: error: narrowing conversion of '10272' from 'uint32_t' {aka 'long unsigned int'} to 'unsigned char' [-Wnarrowing]
844 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(npages + 1);
| ^
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp: In function 'uint32_t process_exec_buf_relay_inline_cmd(uint32_t&, uint32_t&)':
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:939:69: error: no matching function for call to 'cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(uint32_t&)'
939 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(npages);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:14,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id> void cb_release_pages(uint32_t)'
293 | void cb_release_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:939:69: error: narrowing conversion of '10272' from 'uint32_t' {aka 'long unsigned int'} to 'unsigned char' [-Wnarrowing]
939 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(npages);
| ^
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp: In function 'uint32_t process_relay_inline_all(uint32_t, uint32_t, bool)':
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:1215:69: error: no matching function for call to 'cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(uint32_t&)'
1215 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(npages);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:14,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id> void cb_release_pages(uint32_t)'
293 | void cb_release_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:1215:69: error: narrowing conversion of '10272' from 'uint32_t' {aka 'long unsigned int'} to 'unsigned char' [-Wnarrowing]
1215 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(npages);
| ^
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp: In function 'uint32_t relay_cb_get_cmds(uint32_t&, uint32_t&)':
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:1238:48: error: too few arguments to function 'uint32_t get_cb_page(uint32_t&, uint32_t&, uint32_t*, uint32_t*, uint32_t&, uint32_t&) [with long unsigned int cb_base = 110592; long unsigned int cb_blocks = 4; long unsigned int cb_log_page_size = 12; long unsigned int noc_xy = 10256; long unsigned int cb_sem = 0; uint32_t = long unsigned int]'
1238 | rd_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:14,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:382:10: note: declared here
382 | uint32_t get_cb_page(uint32_t& cmd_ptr,
| ^~~~~~~~~~~
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:1264:48: error: too few arguments to function 'uint32_t get_cb_page(uint32_t&, uint32_t&, uint32_t*, uint32_t*, uint32_t&, uint32_t&) [with long unsigned int cb_base = 110592; long unsigned int cb_blocks = 4; long unsigned int cb_log_page_size = 12; long unsigned int noc_xy = 10256; long unsigned int cb_sem = 0; uint32_t = long unsigned int]'
1264 | rd_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:14,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:382:10: note: declared here
382 | uint32_t get_cb_page(uint32_t& cmd_ptr,
| ^~~~~~~~~~~
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp: In function 'void kernel_main_d()':
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:1353:76: error: no matching function for call to 'cb_release_pages<upstream_noc_xy, upstream_cb_sem_id>(uint32_t&)'
1353 | cb_release_pages<upstream_noc_xy, upstream_cb_sem_id>(pages_to_free);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:14,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id> void cb_release_pages(uint32_t)'
293 | void cb_release_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_prefetch.cpp:1353:76: note: couldn't deduce template parameter 'sem_id'
1353 | cb_release_pages<upstream_noc_xy, upstream_cb_sem_id>(pages_to_free);
| ^
Always | FATAL | ncrisc build failed
Always | FATAL | Failed to generate binaries for cq_prefetch TT_THROW @ /tmp/build-via-sdist-ndjr7v6n/metal_libs-0.51.0rc36.dev16+wormhole.b0/tt_metal/jit_build/build.cpp:395: tt::exception
info:
ncrisc build failed
backtrace:
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1b4012e) [0x7f6ebe38d12e]
--- tt::tt_metal::JitBuildState::compile_one(std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, tt::tt_metal::JitBuildSettings const*, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&) const
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1abe373) [0x7f6ebe30b373]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac25ba) [0x7f6ebe30f5ba]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac115d) [0x7f6ebe30e15d]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac0669) [0x7f6ebe30d669]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac04fa) [0x7f6ebe30d4fa]
--- /lib/x86_64-linux-gnu/libpthread.so.0(+0x8609) [0x7f6f7d66c609]
--- /lib/x86_64-linux-gnu/libc.so.6(clone+0x43) [0x7f6f7d7a6353]
BuildKernels | INFO | ncrisc compile failure -- cmd: cd /home/user/tt-metal/.ttnn_runtime_artifacts/built/6160/kernels/cq_dispatch/12444428523747038649/ncrisc/ && /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/third_party/sfpi/compiler/bin/riscv32-unknown-elf-g++ -mwormhole -march=rv32imw -mtune=rvtt-b1 -mabi=ilp32 -std=c++17 -flto -ffast-math -fno-use-cxa-atexit -fno-exceptions -Wall -Werror -Wno-unknown-pragmas -Wno-error=multistatement-macros -Wno-error=parentheses -Wno-error=unused-but-set-variable -Wno-unused-variable -Wno-unused-function -Os -fno-tree-loop-distribute-patterns -DARCH_WORMHOLE -DTENSIX_FIRMWARE -DLOCAL_MEM_EN=0 -DCOMPILE_FOR_NCRISC -DKERNEL_BUILD -DDISPATCH_KERNEL=1 -DDOWNSTREAM_NOC_X=0 -DDOWNSTREAM_NOC_Y=0 -DFD_CORE_TYPE=0 -DMY_NOC_X=2 -DMY_NOC_Y=10 -DUPSTREAM_NOC_INDEX=1 -DUPSTREAM_NOC_X=8 -DUPSTREAM_NOC_Y=1 -DNOC_INDEX=0 -DKERNEL_COMPILE_TIME_ARG_0=110592 -DKERNEL_COMPILE_TIME_ARG_1=12 -DKERNEL_COMPILE_TIME_ARG_2=128 -DKERNEL_COMPILE_TIME_ARG_3=0 -DKERNEL_COMPILE_TIME_ARG_4=1 -DKERNEL_COMPILE_TIME_ARG_5=4 -DKERNEL_COMPILE_TIME_ARG_6=0 -DKERNEL_COMPILE_TIME_ARG_7=0 -DKERNEL_COMPILE_TIME_ARG_8=805310464 -DKERNEL_COMPILE_TIME_ARG_9=268431360 -DKERNEL_COMPILE_TIME_ARG_10=110592 -DKERNEL_COMPILE_TIME_ARG_11=524288 -DKERNEL_COMPILE_TIME_ARG_12=0 -DKERNEL_COMPILE_TIME_ARG_13=0 -DKERNEL_COMPILE_TIME_ARG_14=0 -DKERNEL_COMPILE_TIME_ARG_15=0 -DKERNEL_COMPILE_TIME_ARG_16=0 -DKERNEL_COMPILE_TIME_ARG_17=0 -DKERNEL_COMPILE_TIME_ARG_18=0 -DKERNEL_COMPILE_TIME_ARG_19=56 -DKERNEL_COMPILE_TIME_ARG_20=1 -DKERNEL_COMPILE_TIME_ARG_21=1 -I. -I.. -I/home/user/tt-metal/.ttnn_runtime_artifacts/ -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/include -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/inc -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/inc/debug -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/inc/wormhole -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/inc/wormhole/wormhole_b0_defines -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/inc/wormhole/noc -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/third_party/umd/device/wormhole_b0 -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/ckernels/wormhole_b0/metal/common -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/ckernels/wormhole_b0/metal/llk_io -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/third_party/tt_llk_wormhole_b0/common/inc -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/third_party/tt_llk_wormhole_b0/llk_lib -I /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src -I /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/ckernels/wormhole_b0/metal/common -I /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/ckernels/wormhole_b0/metal/llk_io -c -o ncrisck.o /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp: In function 'void process_write_host_h()':
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:181:62: error: no matching function for call to 'cb_acquire_pages<my_noc_xy, my_dispatch_cb_sem_id, dispatch_cb_log_page_size>(uint32_t&, uint32_t [4], uint32_t&)'
181 | cb_fence, block_next_start_addr, rd_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:272:6: note: candidate: 'template<long unsigned int noc_xy, long unsigned int sem_id> void cb_acquire_pages(uint32_t)'
272 | void cb_acquire_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:272:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:181:62: error: wrong number of template arguments (3, should be 2)
181 | cb_fence, block_next_start_addr, rd_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:301:10: note: candidate: 'uint32_t cb_acquire_pages(uint32_t, uint32_t*, uint32_t, uint32_t&) [with long unsigned int noc_xy = 10272; long unsigned int sem_id = 0; long unsigned int cb_log_page_size = 12; uint32_t = long unsigned int]'
301 | uint32_t cb_acquire_pages(uint32_t cb_fence,
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:301:10: note: candidate expects 4 arguments, 3 provided
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:191:85: error: no matching function for call to 'cb_block_release_pages<upstream_noc_xy, upstream_dispatch_cb_sem_id, dispatch_cb_blocks, dispatch_cb_pages_per_block>(uint32_t [4], uint32_t&)'
191 | dispatch_cb_pages_per_block>(block_noc_writes_to_clear, wr_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:336:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id, long unsigned int cb_blocks, long unsigned int cb_pages_per_block> void cb_block_release_pages(uint32_t*, uint32_t&)'
336 | void cb_block_release_pages(uint32_t block_noc_writes_to_clear[],
| ^~~~~~~~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:336:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:191:85: error: narrowing conversion of '1152' from 'uint32_t' {aka 'long unsigned int'} to 'unsigned char' [-Wnarrowing]
191 | dispatch_cb_pages_per_block>(block_noc_writes_to_clear, wr_block_idx);
| ^
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp: In function 'void relay_to_next_cb(uint32_t, uint32_t)':
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:313:62: error: no matching function for call to 'cb_acquire_pages<my_noc_xy, my_dispatch_cb_sem_id, dispatch_cb_log_page_size>(uint32_t&, uint32_t [4], uint32_t&)'
313 | cb_fence, block_next_start_addr, rd_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:272:6: note: candidate: 'template<long unsigned int noc_xy, long unsigned int sem_id> void cb_acquire_pages(uint32_t)'
272 | void cb_acquire_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:272:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:313:62: error: wrong number of template arguments (3, should be 2)
313 | cb_fence, block_next_start_addr, rd_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:301:10: note: candidate: 'uint32_t cb_acquire_pages(uint32_t, uint32_t*, uint32_t, uint32_t&) [with long unsigned int noc_xy = 10272; long unsigned int sem_id = 0; long unsigned int cb_log_page_size = 12; uint32_t = long unsigned int]'
301 | uint32_t cb_acquire_pages(uint32_t cb_fence,
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:301:10: note: candidate expects 4 arguments, 3 provided
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:322:85: error: no matching function for call to 'cb_block_release_pages<upstream_noc_xy, upstream_dispatch_cb_sem_id, dispatch_cb_blocks, dispatch_cb_pages_per_block>(uint32_t [4], uint32_t&)'
322 | dispatch_cb_pages_per_block>(block_noc_writes_to_clear, wr_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:336:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id, long unsigned int cb_blocks, long unsigned int cb_pages_per_block> void cb_block_release_pages(uint32_t*, uint32_t&)'
336 | void cb_block_release_pages(uint32_t block_noc_writes_to_clear[],
| ^~~~~~~~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:336:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:322:85: error: narrowing conversion of '1152' from 'uint32_t' {aka 'long unsigned int'} to 'unsigned char' [-Wnarrowing]
322 | dispatch_cb_pages_per_block>(block_noc_writes_to_clear, wr_block_idx);
| ^
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:327:68: error: no matching function for call to 'cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(int)'
327 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(1); // XXXX optimize, take all available
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id> void cb_release_pages(uint32_t)'
293 | void cb_release_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:327:68: note: couldn't deduce template parameter 'sem_id'
327 | cb_release_pages<downstream_noc_xy, downstream_cb_sem_id>(1); // XXXX optimize, take all available
| ^
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp: In function 'void process_write_linear(uint32_t)':
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:410:62: error: no matching function for call to 'cb_acquire_pages<my_noc_xy, my_dispatch_cb_sem_id, dispatch_cb_log_page_size>(uint32_t&, uint32_t [4], uint32_t&)'
410 | cb_fence, block_next_start_addr, rd_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:272:6: note: candidate: 'template<long unsigned int noc_xy, long unsigned int sem_id> void cb_acquire_pages(uint32_t)'
272 | void cb_acquire_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:272:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:410:62: error: wrong number of template arguments (3, should be 2)
410 | cb_fence, block_next_start_addr, rd_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:301:10: note: candidate: 'uint32_t cb_acquire_pages(uint32_t, uint32_t*, uint32_t, uint32_t&) [with long unsigned int noc_xy = 10272; long unsigned int sem_id = 0; long unsigned int cb_log_page_size = 12; uint32_t = long unsigned int]'
301 | uint32_t cb_acquire_pages(uint32_t cb_fence,
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:301:10: note: candidate expects 4 arguments, 3 provided
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:419:85: error: no matching function for call to 'cb_block_release_pages<upstream_noc_xy, upstream_dispatch_cb_sem_id, dispatch_cb_blocks, dispatch_cb_pages_per_block>(uint32_t [4], uint32_t&)'
419 | dispatch_cb_pages_per_block>(block_noc_writes_to_clear, wr_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:336:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id, long unsigned int cb_blocks, long unsigned int cb_pages_per_block> void cb_block_release_pages(uint32_t*, uint32_t&)'
336 | void cb_block_release_pages(uint32_t block_noc_writes_to_clear[],
| ^~~~~~~~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:336:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:419:85: error: narrowing conversion of '1152' from 'uint32_t' {aka 'long unsigned int'} to 'unsigned char' [-Wnarrowing]
419 | dispatch_cb_pages_per_block>(block_noc_writes_to_clear, wr_block_idx);
| ^
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp: In function 'void process_write_paged()':
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:495:62: error: no matching function for call to 'cb_acquire_pages<my_noc_xy, my_dispatch_cb_sem_id, dispatch_cb_log_page_size>(uint32_t&, uint32_t [4], uint32_t&)'
495 | cb_fence, block_next_start_addr, rd_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:272:6: note: candidate: 'template<long unsigned int noc_xy, long unsigned int sem_id> void cb_acquire_pages(uint32_t)'
272 | void cb_acquire_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:272:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:495:62: error: wrong number of template arguments (3, should be 2)
495 | cb_fence, block_next_start_addr, rd_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:301:10: note: candidate: 'uint32_t cb_acquire_pages(uint32_t, uint32_t*, uint32_t, uint32_t&) [with long unsigned int noc_xy = 10272; long unsigned int sem_id = 0; long unsigned int cb_log_page_size = 12; uint32_t = long unsigned int]'
301 | uint32_t cb_acquire_pages(uint32_t cb_fence,
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:301:10: note: candidate expects 4 arguments, 3 provided
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:504:85: error: no matching function for call to 'cb_block_release_pages<upstream_noc_xy, upstream_dispatch_cb_sem_id, dispatch_cb_blocks, dispatch_cb_pages_per_block>(uint32_t [4], uint32_t&)'
504 | dispatch_cb_pages_per_block>(block_noc_writes_to_clear, wr_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:336:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id, long unsigned int cb_blocks, long unsigned int cb_pages_per_block> void cb_block_release_pages(uint32_t*, uint32_t&)'
336 | void cb_block_release_pages(uint32_t block_noc_writes_to_clear[],
| ^~~~~~~~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:336:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:504:85: error: narrowing conversion of '1152' from 'uint32_t' {aka 'long unsigned int'} to 'unsigned char' [-Wnarrowing]
504 | dispatch_cb_pages_per_block>(block_noc_writes_to_clear, wr_block_idx);
| ^
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp: In function 'void process_write_packed(uint32_t)':
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:602:62: error: no matching function for call to 'cb_acquire_pages<my_noc_xy, my_dispatch_cb_sem_id, dispatch_cb_log_page_size>(uint32_t&, uint32_t [4], uint32_t&)'
602 | cb_fence, block_next_start_addr, rd_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:272:6: note: candidate: 'template<long unsigned int noc_xy, long unsigned int sem_id> void cb_acquire_pages(uint32_t)'
272 | void cb_acquire_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:272:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:602:62: error: wrong number of template arguments (3, should be 2)
602 | cb_fence, block_next_start_addr, rd_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:301:10: note: candidate: 'uint32_t cb_acquire_pages(uint32_t, uint32_t*, uint32_t, uint32_t&) [with long unsigned int noc_xy = 10272; long unsigned int sem_id = 0; long unsigned int cb_log_page_size = 12; uint32_t = long unsigned int]'
301 | uint32_t cb_acquire_pages(uint32_t cb_fence,
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:301:10: note: candidate expects 4 arguments, 3 provided
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:640:77: error: no matching function for call to 'cb_block_release_pages<upstream_noc_xy, upstream_dispatch_cb_sem_id, dispatch_cb_blocks, dispatch_cb_pages_per_block>(uint32_t [4], uint32_t&)'
640 | dispatch_cb_pages_per_block>(block_noc_writes_to_clear, wr_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:336:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id, long unsigned int cb_blocks, long unsigned int cb_pages_per_block> void cb_block_release_pages(uint32_t*, uint32_t&)'
336 | void cb_block_release_pages(uint32_t block_noc_writes_to_clear[],
| ^~~~~~~~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:336:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:640:77: error: narrowing conversion of '1152' from 'uint32_t' {aka 'long unsigned int'} to 'unsigned char' [-Wnarrowing]
640 | dispatch_cb_pages_per_block>(block_noc_writes_to_clear, wr_block_idx);
| ^
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp: In function 'void process_write_packed_large()':
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:722:66: error: no matching function for call to 'cb_acquire_pages<my_noc_xy, my_dispatch_cb_sem_id, dispatch_cb_log_page_size>(uint32_t&, uint32_t [4], uint32_t&)'
722 | cb_fence, block_next_start_addr, rd_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:272:6: note: candidate: 'template<long unsigned int noc_xy, long unsigned int sem_id> void cb_acquire_pages(uint32_t)'
272 | void cb_acquire_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:272:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:722:66: error: wrong number of template arguments (3, should be 2)
722 | cb_fence, block_next_start_addr, rd_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:301:10: note: candidate: 'uint32_t cb_acquire_pages(uint32_t, uint32_t*, uint32_t, uint32_t&) [with long unsigned int noc_xy = 10272; long unsigned int sem_id = 0; long unsigned int cb_log_page_size = 12; uint32_t = long unsigned int]'
301 | uint32_t cb_acquire_pages(uint32_t cb_fence,
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:301:10: note: candidate expects 4 arguments, 3 provided
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:758:62: error: no matching function for call to 'cb_acquire_pages<my_noc_xy, my_dispatch_cb_sem_id, dispatch_cb_log_page_size>(uint32_t&, uint32_t [4], uint32_t&)'
758 | cb_fence, block_next_start_addr, rd_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:272:6: note: candidate: 'template<long unsigned int noc_xy, long unsigned int sem_id> void cb_acquire_pages(uint32_t)'
272 | void cb_acquire_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:272:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:758:62: error: wrong number of template arguments (3, should be 2)
758 | cb_fence, block_next_start_addr, rd_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:301:10: note: candidate: 'uint32_t cb_acquire_pages(uint32_t, uint32_t*, uint32_t, uint32_t&) [with long unsigned int noc_xy = 10272; long unsigned int sem_id = 0; long unsigned int cb_log_page_size = 12; uint32_t = long unsigned int]'
301 | uint32_t cb_acquire_pages(uint32_t cb_fence,
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:301:10: note: candidate expects 4 arguments, 3 provided
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:767:81: error: no matching function for call to 'cb_block_release_pages<upstream_noc_xy, upstream_dispatch_cb_sem_id, dispatch_cb_blocks, dispatch_cb_pages_per_block>(uint32_t [4], uint32_t&)'
767 | dispatch_cb_pages_per_block>(block_noc_writes_to_clear, wr_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:336:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id, long unsigned int cb_blocks, long unsigned int cb_pages_per_block> void cb_block_release_pages(uint32_t*, uint32_t&)'
336 | void cb_block_release_pages(uint32_t block_noc_writes_to_clear[],
| ^~~~~~~~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:336:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:767:81: error: narrowing conversion of '1152' from 'uint32_t' {aka 'long unsigned int'} to 'unsigned char' [-Wnarrowing]
767 | dispatch_cb_pages_per_block>(block_noc_writes_to_clear, wr_block_idx);
| ^
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp: In function 'void kernel_main()':
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:1040:98: error: too few arguments to function 'uint32_t get_cb_page(uint32_t&, uint32_t&, uint32_t*, uint32_t*, uint32_t&, uint32_t&) [with long unsigned int cb_base = 110592; long unsigned int cb_blocks = 4; long unsigned int cb_log_page_size = 12; long unsigned int noc_xy = 10272; long unsigned int cb_sem = 0; uint32_t = long unsigned int]'
1040 | cmd_ptr, cb_fence, block_noc_writes_to_clear, block_next_start_addr, rd_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:382:10: note: declared here
382 | uint32_t get_cb_page(uint32_t& cmd_ptr,
| ^~~~~~~~~~~
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:1056:81: error: no matching function for call to 'cb_block_release_pages<upstream_noc_xy, upstream_dispatch_cb_sem_id, dispatch_cb_blocks, dispatch_cb_pages_per_block>(uint32_t [4], uint32_t&)'
1056 | dispatch_cb_pages_per_block>(block_noc_writes_to_clear, wr_block_idx);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:336:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id, long unsigned int cb_blocks, long unsigned int cb_pages_per_block> void cb_block_release_pages(uint32_t*, uint32_t&)'
336 | void cb_block_release_pages(uint32_t block_noc_writes_to_clear[],
| ^~~~~~~~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:336:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:1056:81: error: narrowing conversion of '1152' from 'uint32_t' {aka 'long unsigned int'} to 'unsigned char' [-Wnarrowing]
1056 | dispatch_cb_pages_per_block>(block_noc_writes_to_clear, wr_block_idx);
| ^
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:1072:99: error: no matching function for call to 'cb_release_pages<upstream_noc_xy, upstream_dispatch_cb_sem_id>(const uint32_t&)'
1072 | cb_release_pages<upstream_noc_xy, upstream_dispatch_cb_sem_id>(dispatch_cb_pages_per_block);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id> void cb_release_pages(uint32_t)'
293 | void cb_release_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:1072:99: error: narrowing conversion of '1152' from 'uint32_t' {aka 'long unsigned int'} to 'unsigned char' [-Wnarrowing]
1072 | cb_release_pages<upstream_noc_xy, upstream_dispatch_cb_sem_id>(dispatch_cb_pages_per_block);
| ^
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:1076:74: error: no matching function for call to 'cb_release_pages<upstream_noc_xy, upstream_dispatch_cb_sem_id>(uint32_t&)'
1076 | cb_release_pages<upstream_noc_xy, upstream_dispatch_cb_sem_id>(npages);
| ^
In file included from /home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:17,
from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: candidate: 'template<unsigned char noc_idx, long unsigned int noc_xy, long unsigned int sem_id> void cb_release_pages(uint32_t)'
293 | void cb_release_pages(uint32_t n) {
| ^~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/impl/dispatch/kernels/cq_common.hpp:293:6: note: template argument deduction/substitution failed:
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:1076:74: error: narrowing conversion of '1152' from 'uint32_t' {aka 'long unsigned int'} to 'unsigned char' [-Wnarrowing]
1076 | cb_release_pages<upstream_noc_xy, upstream_dispatch_cb_sem_id>(npages);
| ^
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp: In instantiation of 'void process_write_paged() [with bool is_dram = true]':
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:879:43: required from here
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:456:33: error: use of deleted function 'InterleavedAddrGen<true>::InterleavedAddrGen()'
456 | InterleavedAddrGen<is_dram> addr_gen;
| ^~~~~~~~
In file included from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/tools/profiler/kernel_profiler.hpp:11,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:16:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/inc/dataflow_api.h:840:8: note: 'InterleavedAddrGen<true>::InterleavedAddrGen()' is implicitly deleted because the default definition would be ill-formed:
840 | struct InterleavedAddrGen {
| ^~~~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/inc/dataflow_api.h:840:8: error: uninitialized const member in 'struct InterleavedAddrGen<true>'
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/inc/dataflow_api.h:842:20: note: 'const uint32_t InterleavedAddrGen<true>::page_size' should be initialized
842 | const uint32_t page_size; // Num bytes in page.
| ^~~~~~~~~
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:458:24: error: assignment of read-only member 'InterleavedAddrGen<true>::page_size'
458 | addr_gen.page_size = page_size;
| ~~~~~~~~~~~~~~~~~~~^~~~~~~~~~~
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp: In instantiation of 'void process_write_paged() [with bool is_dram = false]':
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:881:44: required from here
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:456:33: error: use of deleted function 'InterleavedAddrGen<false>::InterleavedAddrGen()'
456 | InterleavedAddrGen<is_dram> addr_gen;
| ^~~~~~~~
In file included from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/tools/profiler/kernel_profiler.hpp:11,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:16:
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/inc/dataflow_api.h:840:8: note: 'InterleavedAddrGen<false>::InterleavedAddrGen()' is implicitly deleted because the default definition would be ill-formed:
840 | struct InterleavedAddrGen {
| ^~~~~~~~~~~~~~~~~~
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/inc/dataflow_api.h:840:8: error: uninitialized const member in 'struct InterleavedAddrGen<false>'
/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/inc/dataflow_api.h:842:20: note: 'const uint32_t InterleavedAddrGen<false>::page_size' should be initialized
842 | const uint32_t page_size; // Num bytes in page.
| ^~~~~~~~~
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/ncrisck.cc:21:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/cq_dispatch.cpp:458:24: error: assignment of read-only member 'InterleavedAddrGen<false>::page_size'
458 | addr_gen.page_size = page_size;
| ~~~~~~~~~~~~~~~~~~~^~~~~~~~~~~
Always | FATAL | ncrisc build failed
Always | FATAL | Failed to generate binaries for cq_dispatch TT_THROW @ /tmp/build-via-sdist-ndjr7v6n/metal_libs-0.51.0rc36.dev16+wormhole.b0/tt_metal/jit_build/build.cpp:395: tt::exception
info:
ncrisc build failed
backtrace:
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1b4012e) [0x7f6ebe38d12e]
--- tt::tt_metal::JitBuildState::compile_one(std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, tt::tt_metal::JitBuildSettings const*, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&) const
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1abe373) [0x7f6ebe30b373]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac25ba) [0x7f6ebe30f5ba]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac115d) [0x7f6ebe30e15d]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac0669) [0x7f6ebe30d669]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac04fa) [0x7f6ebe30d4fa]
--- /lib/x86_64-linux-gnu/libpthread.so.0(+0x8609) [0x7f6f7d66c609]
--- /lib/x86_64-linux-gnu/libc.so.6(clone+0x43) [0x7f6f7d7a6353]
libc++abi: terminating due to uncaught exception of type std::runtime_error: TT_THROW @ /tmp/build-via-sdist-ndjr7v6n/metal_libs-0.51.0rc36.dev16+wormhole.b0/tt_metal/impl/program/program.cpp:39: tt::exception
info:
Failed to generate binaries for cq_dispatch TT_THROW @ /tmp/build-via-sdist-ndjr7v6n/metal_libs-0.51.0rc36.dev16+wormhole.b0/tt_metal/jit_build/build.cpp:395: tt::exception
info:
ncrisc build failed
backtrace:
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1b4012e) [0x7f6ebe38d12e]
--- tt::tt_metal::JitBuildState::compile_one(std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, tt::tt_metal::JitBuildSettings const*, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&) const
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1abe373) [0x7f6ebe30b373]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac25ba) [0x7f6ebe30f5ba]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac115d) [0x7f6ebe30e15d]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac0669) [0x7f6ebe30d669]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac04fa) [0x7f6ebe30d4fa]
--- /lib/x86_64-linux-gnu/libpthread.so.0(+0x8609) [0x7f6f7d66c609]
--- /lib/x86_64-linux-gnu/libc.so.6(clone+0x43) [0x7f6f7d7a6353]
backtrace:
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ab647e) [0x7f6ebe30347e]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1abe373) [0x7f6ebe30b373]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac25ba) [0x7f6ebe30f5ba]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac115d) [0x7f6ebe30e15d]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac0669) [0x7f6ebe30d669]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac04fa) [0x7f6ebe30d4fa]
--- /lib/x86_64-linux-gnu/libpthread.so.0(+0x8609) [0x7f6f7d66c609]
--- /lib/x86_64-linux-gnu/libc.so.6(clone+0x43) [0x7f6f7d7a6353]
Fatal Python error: Aborted
Current thread 0x00007f6f7d4bb740 (most recent call first):
File "/usr/local/lib/python3.8/dist-packages/ttnn/multi_device.py", line 149 in open_device_mesh
File "/home/user/tt-metal/conftest.py", line 243 in t3k_device_mesh
File "/usr/local/lib/python3.8/dist-packages/_pytest/fixtures.py", line 901 in call_fixture_func
File "/usr/local/lib/python3.8/dist-packages/_pytest/fixtures.py", line 1129 in pytest_fixture_setup
File "/usr/local/lib/python3.8/dist-packages/pluggy/_callers.py", line 103 in _multicall
File "/usr/local/lib/python3.8/dist-packages/pluggy/_manager.py", line 120 in _hookexec
File "/usr/local/lib/python3.8/dist-packages/pluggy/_hooks.py", line 513 in __call__
File "/usr/local/lib/python3.8/dist-packages/_pytest/fixtures.py", line 1075 in execute
File "/usr/local/lib/python3.8/dist-packages/_pytest/fixtures.py", line 677 in _compute_fixture_value
File "/usr/local/lib/python3.8/dist-packages/_pytest/fixtures.py", line 591 in _get_active_fixturedef
File "/usr/local/lib/python3.8/dist-packages/_pytest/fixtures.py", line 569 in getfixturevalue
File "/usr/local/lib/python3.8/dist-packages/_pytest/fixtures.py", line 550 in _fillfixtures
File "/usr/local/lib/python3.8/dist-packages/_pytest/python.py", line 1792 in setup
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 492 in setup
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 155 in pytest_runtest_setup
File "/usr/local/lib/python3.8/dist-packages/pluggy/_callers.py", line 103 in _multicall
File "/usr/local/lib/python3.8/dist-packages/pluggy/_manager.py", line 120 in _hookexec
File "/usr/local/lib/python3.8/dist-packages/pluggy/_hooks.py", line 513 in __call__
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 260 in <lambda>
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 339 in from_call
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 259 in call_runtest_hook
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 220 in call_and_report
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 125 in runtestprotocol
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 112 in pytest_runtest_protocol
File "/usr/local/lib/python3.8/dist-packages/pluggy/_callers.py", line 103 in _multicall
File "/usr/local/lib/python3.8/dist-packages/pluggy/_manager.py", line 120 in _hookexec
File "/usr/local/lib/python3.8/dist-packages/pluggy/_hooks.py", line 513 in __call__
File "/usr/local/lib/python3.8/dist-packages/_pytest/main.py", line 349 in pytest_runtestloop
File "/usr/local/lib/python3.8/dist-packages/pluggy/_callers.py", line 103 in _multicall
File "/usr/local/lib/python3.8/dist-packages/pluggy/_manager.py", line 120 in _hookexec
File "/usr/local/lib/python3.8/dist-packages/pluggy/_hooks.py", line 513 in __call__
File "/usr/local/lib/python3.8/dist-packages/_pytest/main.py", line 324 in _main
File "/usr/local/lib/python3.8/dist-packages/_pytest/main.py", line 270 in wrap_session
File "/usr/local/lib/python3.8/dist-packages/_pytest/main.py", line 317 in pytest_cmdline_main
File "/usr/local/lib/python3.8/dist-packages/pluggy/_callers.py", line 103 in _multicall
File "/usr/local/lib/python3.8/dist-packages/pluggy/_manager.py", line 120 in _hookexec
File "/usr/local/lib/python3.8/dist-packages/pluggy/_hooks.py", line 513 in __call__
File "/usr/local/lib/python3.8/dist-packages/_pytest/config/__init__.py", line 167 in main
File "/usr/local/lib/python3.8/dist-packages/_pytest/config/__init__.py", line 190 in console_main
File "/usr/local/bin/pytest", line 8 in <module>
Aborted (core dumped)
As noted by @tt-rkim, my source commit was not right. I switched to the right commit (37c13738e274d101aa29642af9c5c85ed20fcfd1) and followed the steps below:
Installing whl (Dockerfile steps)
RUN curl -L \
-H "Accept: application/vnd.github+json" \
-H "Authorization: Bearer ${GITHUB_TOKEN}" \
-H "X-GitHub-Api-Version: 2022-11-28" \
--output /tmp/artifact.zip \
https://api.github.com/repos/tenstorrent/tt-metal/actions/artifacts/1854934825/zip
# Unzip the artifact
RUN unzip /tmp/artifact.zip -d /tmp/artifact
# Find and install the .whl file inside the unzipped directory
RUN find /tmp/artifact -name "*.whl" -exec pip install {} \;
Cloning Repo git clone https://github.com/tenstorrent/tt-metal.git --recurse-submodules cd tt-metal git fetch origin 37c13738e274d101aa29642af9c5c85ed20fcfd1 git checkout https://github.com/tenstorrent/tt-metal/commit/37c13738e274d101aa29642af9c5c85ed20fcfd1 git submodule update --init --recursive git submodule foreach 'git lfs fetch --all && git lfs pull’
export PYTHONPATH=$(pwd) pip install -r tt_metal/python_env/requirements-dev.txt
Setup Mixtral export TT_METAL_ASYNC_DEVICE_QUEUE=1 export WH_ARCH_YAML=wormhole_b0_80_arch_eth_dispatch.yaml
export MIXTRAL_CKPT_DIR=/mnt/kanishk-mldata/mixtral/mixtral_consolidated/Mixtral-8x7B-v0.1/repacked_weights_dir export MIXTRAL_TOKENIZER_PATH=/mnt/kanishk-mldata/mixtral/mixtral_consolidated/Mixtral-8x7B-v0.1 export MIXTRAL_CACHE_PATH=/mnt/kanishk-mldata/mixtral/mixtral_consolidated/Mixtral-8x7B-v0.1/weights_cache_dir
Run Mixtral pytest -svv models/demos/t3000/mixtral8x7b/demo/demo.py::test_mixtral8x7b_demo[wormhole_b0-True-general_weights]
Error log:
user@llama31-70b-4837c6bc-deployment-6d55668c84-b42xd:~/tt-metal$ pytest -svv models/demos/t3000/mixtral8x7b/demo/demo.py::test_mixtral8x7b_demo[wormhole_b0-True-general_weights]
2024-08-27 00:04:19.314 | DEBUG | ttnn.library_tweaks:prepare_dir_as_metal_home:54 - Existing installation of 0.51.0rc36.dev16+wormhole.b0 detected
2024-08-27 00:04:19.412 | DEBUG | ttnn:<module>:82 - Initial ttnn.CONFIG:
Config{cache_path=/home/user/.cache/ttnn,model_cache_path=/home/user/.cache/ttnn/models,tmp_dir=/tmp/ttnn,enable_model_cache=false,enable_fast_runtime_mode=true,throw_exception_on_fallback=false,enable_logging=false,enable_graph_report=false,enable_detailed_buffer_report=false,enable_detailed_tensor_report=false,enable_comparison_mode=false,comparison_mode_pcc=0.9999,root_report_path=generated/ttnn/reports,report_name=std::nullopt,std::nullopt}
2024-08-27 00:04:19.485 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.pearson_correlation_coefficient be migrated to C++?
2024-08-27 00:04:19.486 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.Conv1d be migrated to C++?
2024-08-27 00:04:19.489 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.conv2d be migrated to C++?
2024-08-27 00:04:19.489 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.reshape be migrated to C++?
2024-08-27 00:04:19.489 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.unsqueeze_to_4D be migrated to C++?
2024-08-27 00:04:19.490 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.squeeze be migrated to C++?
2024-08-27 00:04:19.490 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.from_torch be migrated to C++?
2024-08-27 00:04:19.490 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.to_torch be migrated to C++?
2024-08-27 00:04:19.490 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.to_device be migrated to C++?
2024-08-27 00:04:19.490 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.from_device be migrated to C++?
2024-08-27 00:04:19.490 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.allocate_tensor_on_device be migrated to C++?
2024-08-27 00:04:19.490 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.copy_host_to_device_tensor be migrated to C++?
2024-08-27 00:04:19.490 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.deallocate be migrated to C++?
2024-08-27 00:04:19.490 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.reallocate be migrated to C++?
2024-08-27 00:04:19.491 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.load_tensor be migrated to C++?
2024-08-27 00:04:19.491 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.dump_tensor be migrated to C++?
2024-08-27 00:04:19.491 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.as_tensor be migrated to C++?
2024-08-27 00:04:19.492 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.max_pool2d be migrated to C++?
2024-08-27 00:04:19.492 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.max_pool2d_legacy be migrated to C++?
2024-08-27 00:04:19.492 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.global_avg_pool2d be migrated to C++?
2024-08-27 00:04:19.492 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.avg_pool2d be migrated to C++?
2024-08-27 00:04:19.495 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.conv2d be migrated to C++?
2024-08-27 00:04:19.495 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.max_pool2d be migrated to C++?
2024-08-27 00:04:19.495 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.max_pool2d_legacy be migrated to C++?
2024-08-27 00:04:19.496 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.global_avg_pool2d be migrated to C++?
2024-08-27 00:04:19.496 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.avg_pool2d be migrated to C++?
2024-08-27 00:04:19.496 | WARNING | ttnn.decorators:operation_decorator:776 - Should ttnn.Conv1d be migrated to C++?
==================================================================== test session starts =====================================================================
platform linux -- Python 3.8.10, pytest-7.2.2, pluggy-1.5.0 -- /usr/bin/python3
cachedir: .pytest_cache
rootdir: /home/user/tt-metal, configfile: pytest.ini
plugins: anyio-4.4.0, dash-2.15.0, split-0.8.2, xdist-3.6.1, timeout-2.2.0
timeout: 300.0s
timeout method: signal
timeout func_only: False
collected 1 item
models/demos/t3000/mixtral8x7b/demo/demo.py::test_mixtral8x7b_demo[wormhole_b0-True-general_weights] Device | INFO | Opening user mode device driver
Detecting chips (found 8)
2024-08-27 00:04:20.836 | INFO | SiliconDriver - Detected 4 PCI devices : [0, 1, 2, 3]
2024-08-27 00:04:20.913 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 0)
2024-08-27 00:04:20.914 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 4)
2024-08-27 00:04:20.917 | INFO | SiliconDriver - Detected 4 PCI devices : [0, 1, 2, 3]
2024-08-27 00:04:20.932 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 1)
2024-08-27 00:04:20.932 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 5)
2024-08-27 00:04:20.936 | INFO | SiliconDriver - Detected 4 PCI devices : [0, 1, 2, 3]
2024-08-27 00:04:20.950 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 2)
2024-08-27 00:04:20.951 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 6)
2024-08-27 00:04:20.954 | INFO | SiliconDriver - Detected 4 PCI devices : [0, 1, 2, 3]
2024-08-27 00:04:20.969 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 3)
2024-08-27 00:04:20.969 | INFO | SiliconDriver - Software version 6.0.0, Ethernet FW version 6.9.0 (Device 7)
Metal | INFO | Initializing device 0. Program cache is NOT enabled
Metal | INFO | AI CLK for device 0 is: 1000 MHz
Metal | INFO | Initializing device 4. Program cache is NOT enabled
Metal | INFO | AI CLK for device 4 is: 1000 MHz
Metal | INFO | Initializing device 1. Program cache is NOT enabled
Metal | INFO | AI CLK for device 1 is: 1000 MHz
Metal | INFO | Initializing device 5. Program cache is NOT enabled
Metal | INFO | AI CLK for device 5 is: 1000 MHz
Metal | INFO | Initializing device 2. Program cache is NOT enabled
Metal | INFO | AI CLK for device 2 is: 1000 MHz
Metal | INFO | Initializing device 6. Program cache is NOT enabled
Metal | INFO | AI CLK for device 6 is: 1000 MHz
Metal | INFO | Initializing device 3. Program cache is NOT enabled
Metal | INFO | AI CLK for device 3 is: 1000 MHz
Metal | INFO | Initializing device 7. Program cache is NOT enabled
Metal | INFO | AI CLK for device 7 is: 1000 MHz
Metal | INFO | MMIO Device 0 : Tunnel 0 : Device 0
Metal | INFO | MMIO Device 0 : Tunnel 0 : Device 4
BuildKernels | INFO | idle_erisc compile failure -- cmd: cd /home/user/tt-metal/.ttnn_runtime_artifacts/built/6160/kernels/packet_demux/12969583511883273621/idle_erisc/ && /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/third_party/sfpi/compiler/bin/riscv32-unknown-elf-g++ -mwormhole -march=rv32imw -mtune=rvtt-b1 -mabi=ilp32 -std=c++17 -flto -ffast-math -fno-use-cxa-atexit -fno-exceptions -Wall -Werror -Wno-unknown-pragmas -Wno-error=multistatement-macros -Wno-error=parentheses -Wno-error=unused-but-set-variable -Wno-unused-variable -Wno-unused-function -Os -fno-tree-loop-distribute-patterns -DARCH_WORMHOLE -DTENSIX_FIRMWARE -DLOCAL_MEM_EN=0 -DCOMPILE_FOR_IDLE_ERISC -DERISC -DRISC_B0_HW -DKERNEL_BUILD -DDISPATCH_KERNEL=1 -DDOWNSTREAM_NOC_X=0 -DDOWNSTREAM_NOC_Y=0 -DFD_CORE_TYPE=2 -DMY_NOC_X=0 -DMY_NOC_Y=0 -DSKIP_NOC_LOGGING=1 -DUPSTREAM_NOC_INDEX=0 -DUPSTREAM_NOC_X=0 -DUPSTREAM_NOC_Y=0 -DNOC_INDEX=0 -DKERNEL_COMPILE_TIME_ARG_0=177 -DKERNEL_COMPILE_TIME_ARG_1=6710 -DKERNEL_COMPILE_TIME_ARG_2=4096 -DKERNEL_COMPILE_TIME_ARG_3=1 -DKERNEL_COMPILE_TIME_ARG_4=1544 -DKERNEL_COMPILE_TIME_ARG_5=0 -DKERNEL_COMPILE_TIME_ARG_6=0 -DKERNEL_COMPILE_TIME_ARG_7=0 -DKERNEL_COMPILE_TIME_ARG_8=6912 -DKERNEL_COMPILE_TIME_ARG_9=8192 -DKERNEL_COMPILE_TIME_ARG_10=0 -DKERNEL_COMPILE_TIME_ARG_11=0 -DKERNEL_COMPILE_TIME_ARG_12=0 -DKERNEL_COMPILE_TIME_ARG_13=0 -DKERNEL_COMPILE_TIME_ARG_14=0 -DKERNEL_COMPILE_TIME_ARG_15=0 -DKERNEL_COMPILE_TIME_ARG_16=1 -DKERNEL_COMPILE_TIME_ARG_17=0 -DKERNEL_COMPILE_TIME_ARG_18=2 -DKERNEL_COMPILE_TIME_ARG_19=0 -DKERNEL_COMPILE_TIME_ARG_20=0 -DKERNEL_COMPILE_TIME_ARG_21=84 -DKERNEL_COMPILE_TIME_ARG_22=0 -DKERNEL_COMPILE_TIME_ARG_23=0 -DKERNEL_COMPILE_TIME_ARG_24=0 -DKERNEL_COMPILE_TIME_ARG_25=1 -DKERNEL_COMPILE_TIME_ARG_26=268 -DKERNEL_COMPILE_TIME_ARG_27=0 -DKERNEL_COMPILE_TIME_ARG_28=0 -DKERNEL_COMPILE_TIME_ARG_29=0 -I. -I.. -I/home/user/tt-metal/.ttnn_runtime_artifacts/ -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/include -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/inc -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/inc/debug -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/inc/wormhole -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/inc/wormhole/wormhole_b0_defines -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/inc/wormhole/noc -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/third_party/umd/device/wormhole_b0 -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/ckernels/wormhole_b0/metal/common -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/ckernels/wormhole_b0/metal/llk_io -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/third_party/tt_llk_wormhole_b0/common/inc -I/home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/third_party/tt_llk_wormhole_b0/llk_lib -I /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/ckernels/wormhole_b0/metal/common -I /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/ckernels/wormhole_b0/metal/llk_io -I /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src -c -o idle_erisck.o /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/idle_erisck.cc
In file included from ../kernel_includes.hpp:1,
from /home/user/tt-metal/.ttnn_runtime_artifacts/tt_metal/hw/firmware/src/idle_erisck.cc:23:
/home/user/tt-metal/tt_metal/impl/dispatch/kernels/packet_demux.cpp:9:10: fatal error: tests/tt_metal/tt_metal/perf_microbenchmark/routing/kernels/traffic_gen.hpp: No such file or directory
9 | #include "tests/tt_metal/tt_metal/perf_microbenchmark/routing/kernels/traffic_gen.hpp"
| ^~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~
compilation terminated.
Always | FATAL | idle_erisc build failed
Always | FATAL | Failed to generate binaries for packet_demux TT_THROW @ /tmp/build-via-sdist-ndjr7v6n/metal_libs-0.51.0rc36.dev16+wormhole.b0/tt_metal/jit_build/build.cpp:395: tt::exception
info:
idle_erisc build failed
backtrace:
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1b4012e) [0x7f8e76c1c12e]
--- tt::tt_metal::JitBuildState::compile_one(std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, tt::tt_metal::JitBuildSettings const*, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&) const
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1abe373) [0x7f8e76b9a373]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac25ba) [0x7f8e76b9e5ba]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac115d) [0x7f8e76b9d15d]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac0669) [0x7f8e76b9c669]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac04fa) [0x7f8e76b9c4fa]
--- /lib/x86_64-linux-gnu/libpthread.so.0(+0x8609) [0x7f8f35efb609]
--- /lib/x86_64-linux-gnu/libc.so.6(clone+0x43) [0x7f8f36035353]
libc++abi: terminating due to uncaught exception of type std::runtime_error: TT_THROW @ /tmp/build-via-sdist-ndjr7v6n/metal_libs-0.51.0rc36.dev16+wormhole.b0/tt_metal/impl/program/program.cpp:39: tt::exception
info:
Failed to generate binaries for packet_demux TT_THROW @ /tmp/build-via-sdist-ndjr7v6n/metal_libs-0.51.0rc36.dev16+wormhole.b0/tt_metal/jit_build/build.cpp:395: tt::exception
info:
idle_erisc build failed
backtrace:
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1b4012e) [0x7f8e76c1c12e]
--- tt::tt_metal::JitBuildState::compile_one(std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, tt::tt_metal::JitBuildSettings const*, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&, std::__1::basic_string<char, std::__1::char_traits<char>, std::__1::allocator<char>> const&) const
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1abe373) [0x7f8e76b9a373]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac25ba) [0x7f8e76b9e5ba]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac115d) [0x7f8e76b9d15d]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac0669) [0x7f8e76b9c669]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac04fa) [0x7f8e76b9c4fa]
--- /lib/x86_64-linux-gnu/libpthread.so.0(+0x8609) [0x7f8f35efb609]
--- /lib/x86_64-linux-gnu/libc.so.6(clone+0x43) [0x7f8f36035353]
backtrace:
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ab647e) [0x7f8e76b9247e]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1abe373) [0x7f8e76b9a373]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac25ba) [0x7f8e76b9e5ba]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac115d) [0x7f8e76b9d15d]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac0669) [0x7f8e76b9c669]
--- /usr/local/lib/python3.8/dist-packages/ttnn/_ttnn.cpython-38-x86_64-linux-gnu.so(+0x1ac04fa) [0x7f8e76b9c4fa]
--- /lib/x86_64-linux-gnu/libpthread.so.0(+0x8609) [0x7f8f35efb609]
--- /lib/x86_64-linux-gnu/libc.so.6(clone+0x43) [0x7f8f36035353]
Fatal Python error: Aborted
Current thread 0x00007f8f35d4a740 (most recent call first):
File "/usr/local/lib/python3.8/dist-packages/ttnn/multi_device.py", line 149 in open_device_mesh
File "/home/user/tt-metal/conftest.py", line 270 in t3k_device_mesh
File "/usr/local/lib/python3.8/dist-packages/_pytest/fixtures.py", line 901 in call_fixture_func
File "/usr/local/lib/python3.8/dist-packages/_pytest/fixtures.py", line 1129 in pytest_fixture_setup
File "/usr/local/lib/python3.8/dist-packages/pluggy/_callers.py", line 103 in _multicall
File "/usr/local/lib/python3.8/dist-packages/pluggy/_manager.py", line 120 in _hookexec
File "/usr/local/lib/python3.8/dist-packages/pluggy/_hooks.py", line 513 in __call__
File "/usr/local/lib/python3.8/dist-packages/_pytest/fixtures.py", line 1075 in execute
File "/usr/local/lib/python3.8/dist-packages/_pytest/fixtures.py", line 677 in _compute_fixture_value
File "/usr/local/lib/python3.8/dist-packages/_pytest/fixtures.py", line 591 in _get_active_fixturedef
File "/usr/local/lib/python3.8/dist-packages/_pytest/fixtures.py", line 569 in getfixturevalue
File "/usr/local/lib/python3.8/dist-packages/_pytest/fixtures.py", line 550 in _fillfixtures
File "/usr/local/lib/python3.8/dist-packages/_pytest/python.py", line 1792 in setup
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 492 in setup
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 155 in pytest_runtest_setup
File "/usr/local/lib/python3.8/dist-packages/pluggy/_callers.py", line 103 in _multicall
File "/usr/local/lib/python3.8/dist-packages/pluggy/_manager.py", line 120 in _hookexec
File "/usr/local/lib/python3.8/dist-packages/pluggy/_hooks.py", line 513 in __call__
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 260 in <lambda>
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 339 in from_call
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 259 in call_runtest_hook
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 220 in call_and_report
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 125 in runtestprotocol
File "/usr/local/lib/python3.8/dist-packages/_pytest/runner.py", line 112 in pytest_runtest_protocol
File "/usr/local/lib/python3.8/dist-packages/pluggy/_callers.py", line 103 in _multicall
File "/usr/local/lib/python3.8/dist-packages/pluggy/_manager.py", line 120 in _hookexec
File "/usr/local/lib/python3.8/dist-packages/pluggy/_hooks.py", line 513 in __call__
File "/usr/local/lib/python3.8/dist-packages/_pytest/main.py", line 349 in pytest_runtestloop
File "/usr/local/lib/python3.8/dist-packages/pluggy/_callers.py", line 103 in _multicall
File "/usr/local/lib/python3.8/dist-packages/pluggy/_manager.py", line 120 in _hookexec
File "/usr/local/lib/python3.8/dist-packages/pluggy/_hooks.py", line 513 in __call__
File "/usr/local/lib/python3.8/dist-packages/_pytest/main.py", line 324 in _main
File "/usr/local/lib/python3.8/dist-packages/_pytest/main.py", line 270 in wrap_session
File "/usr/local/lib/python3.8/dist-packages/_pytest/main.py", line 317 in pytest_cmdline_main
File "/usr/local/lib/python3.8/dist-packages/pluggy/_callers.py", line 103 in _multicall
File "/usr/local/lib/python3.8/dist-packages/pluggy/_manager.py", line 120 in _hookexec
File "/usr/local/lib/python3.8/dist-packages/pluggy/_hooks.py", line 513 in __call__
File "/usr/local/lib/python3.8/dist-packages/_pytest/config/__init__.py", line 167 in main
File "/usr/local/lib/python3.8/dist-packages/_pytest/config/__init__.py", line 190 in console_main
File "/usr/local/bin/pytest", line 8 in <module>
Aborted (core dumped)
Looks like there was another kernel importing from test dir 😞 . Will remove.
I've merged the fix (83772d0). Took a look through the remaining dispatch kernels and doesn't seem like there are any test includes left
I will rebase and re-generate a wheel. Please ensure to checkout the appropriate commit when you try again. @ksriv001
This should be down with an hour: https://github.com/tenstorrent/tt-metal/actions/runs/10579812821
Please use this commit to checkout the repo and this workflow to download the wheel. @ksriv001
Thank you for helping in solving this @tt-aho @tt-rkim After using the following whl asset and commit, the issue went away!
53bbd7346e7647047cf333fa851a59eda6c977cc
pytest -svv models/demos/t3000/mixtral8x7b/demo/demo.py::test_mixtral8x7b_demo[wormhole_b0-True-general]
This is absolutely great!
Are there other things you would like to try, or should we close this issue?
Also, @asrinivasanTT what about your other issue?
@tt-rkim Yes, please go ahead and close the issue. Thanks. Also, lets continue the conversation about the other ticket here: https://github.com/tenstorrent/tt-metal/issues/11854
Description When attempting to run tt-metal demos (and the ops test here from the wheel downloaded from GH assets, it throws an error (screenshot provided). However, running the same demos (and ops test) using the wheel built from source in the same environment works without any issues.
To Reproduce Steps to reproduce the behavior: Though we tested it in a pod in tt-cloud (virtual llm-box), this issue can be reproduced on bare-metal llm-boxes too.
git clone https://github.com/tenstorrent/tt-metal.git --recurse-submodules
cd tt-metal
git submodule foreach 'git lfs fetch --all && git lfs pull'
export TT_METAL_VERSION=v0.50.0
export ARCH_NAME=wormhole_b0
pip install https://github.com/tenstorrent/tt-metal/releases/download/${TT_METAL_VERSION}/metal_libs-$(echo ${TT_METAL_VERSION} | sed -e 's/[v-]//g')+$(echo ${ARCH_NAME} | sed -e 's/_/./g')-cp38-cp38-linux_x86_64.whl
python3 test.py
Error Trace
Case 2: Mixtral 7B on 8 cards (mesh config) Test Also attaching the error trace from running Mixtral test inside a venv having tt-metal installed from the same whls: Steps:
export TT_METAL_ASYNC_DEVICE_QUEUE=1
export WH_ARCH_YAML=wormhole_b0_80_arch_eth_dispatch.yaml
export MIXTRAL_CKPT_DIR=/mnt/kanishk-mldata/mixtral/mixtral_consolidated/Mixtral-8x7B-v0.1/repacked_weights_dir
export MIXTRAL_TOKENIZER_PATH=/mnt/kanishk-mldata/mixtral/mixtral_consolidated/Mixtral-8x7B-v0.1
export MIXTRAL_CACHE_PATH=/mnt/kanishk-mldata/mixtral/mixtral_consolidated/Mixtral-8x7B-v0.1/weights_cache_dir
Cache weights:
pytest -svv models/demos/t3000/mixtral8x7b/tests/test_mixtral_model.py::test_mixtral_model_inference[wormhole_b0-True-1-32-output]
Run demo:
pytest -svv models/demos/t3000/mixtral8x7b/demo/demo.py::test_mixtral8x7b_demo[wormhole_b0-True-general_weights]
Error Trace
Environment information:
40a50a0cbc7df7bcfb64e2aa35f297ccc0ed83bf