From eb6bb0b92dfaec0beed5610dc69225ac56dee411 Mon Sep 17 00:00:00 2001 From: Vincent Tang Date: Tue, 2 Jul 2024 22:28:06 +0000 Subject: [PATCH] #11099: VCS simulator setup - bump umd to pick up simulator infra changes - soc + core descriptors for VCS simulator - CPM package for Google flatbuffer, NNG and libuv - apptainer .def for `aus` machines - TT_METAL_SIMULATOR_EN env var to target VCS simulator - remove everything versim related --- CMakeLists.txt | 15 +++-- cmake/dependencies.cmake | 35 ++++++++++ cmake/umd_device.cmake | 26 ++++++- .../perf_microbenchmark/common/util.hpp | 6 +- tt_metal/common/base.hpp | 6 +- tt_metal/common/core_descriptor.hpp | 17 ++--- tt_metal/common/test_common.hpp | 11 ++- .../blackhole_simulation_1x2_arch.yaml | 34 ++++++++++ .../hw/inc/blackhole/dev_mem_map_versim.h | 11 --- .../hw/inc/grayskull/dev_mem_map_versim.h | 11 --- tt_metal/hw/inc/grayskull/tensix.h | 1 - tt_metal/hw/inc/wormhole/dev_mem_map_versim.h | 11 --- tt_metal/hw/inc/wormhole/tensix.h | 1 - tt_metal/impl/device/device.cpp | 2 - tt_metal/llrt/llrt.hpp | 2 +- tt_metal/llrt/tlb_config.hpp | 2 +- tt_metal/llrt/tt_cluster.cpp | 65 +++++++++--------- tt_metal/llrt/tt_cluster.hpp | 2 +- ...aml => blackhole_simulation_1x2_arch.yaml} | 19 ++---- .../blackhole_versim_1x1_arch.yaml | 67 ------------------- .../wormhole_b0_versim_1x1_arch.yaml | 65 ------------------ tt_metal/third_party/umd | 2 +- tt_metal/tt_metal.cpp | 8 --- ttnn/CMakeLists.txt | 2 +- 24 files changed, 158 insertions(+), 263 deletions(-) create mode 100644 tt_metal/core_descriptors/blackhole_simulation_1x2_arch.yaml delete mode 100644 tt_metal/hw/inc/blackhole/dev_mem_map_versim.h delete mode 100644 tt_metal/hw/inc/grayskull/dev_mem_map_versim.h delete mode 100644 tt_metal/hw/inc/wormhole/dev_mem_map_versim.h rename tt_metal/soc_descriptors/{grayskull_versim_1x1_arch.yaml => blackhole_simulation_1x2_arch.yaml} (64%) delete mode 100644 tt_metal/soc_descriptors/blackhole_versim_1x1_arch.yaml delete mode 100644 tt_metal/soc_descriptors/wormhole_b0_versim_1x1_arch.yaml diff --git a/CMakeLists.txt b/CMakeLists.txt index f26ba7d0fac..9f42a8512ae 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -69,7 +69,7 @@ option(WITH_PYTHON_BINDINGS "Enables build of python bindings" ON) message(STATUS "Build Python bindings: ${WITH_PYTHON_BINDINGS}") option(ENABLE_CODE_TIMERS "Enable code timers" OFF) -option(TT_METAL_VERSIM_DISABLED "Disable TT_METAL_VERSIM" ON) + option(ENABLE_TRACY "Enable Tracy Profiling" OFF) @@ -121,7 +121,13 @@ endif() ############################################################################################################################ add_library(stdlib INTERFACE) if(CMAKE_CXX_COMPILER_ID MATCHES "Clang") - target_link_libraries(stdlib INTERFACE c++ c++abi) + find_library(LIBC++ c++) + find_library(LIBC++ABI c++abi) + if(NOT LIBC++ OR NOT LIBC++ABI) + message(FATAL_ERROR "libc++ or libc++abi not found. Make sure you have libc++ and libc++abi installed and in your PATH") + endif() + + target_link_libraries(stdlib INTERFACE ${LIBC++} ${LIBC++ABI}) target_compile_options(stdlib INTERFACE -stdlib=libc++) else() target_link_libraries(stdlib INTERFACE stdc++) @@ -147,9 +153,6 @@ add_library(compiler_flags INTERFACE) target_link_libraries(compiler_flags INTERFACE compiler_warnings stdlib) target_compile_options(compiler_flags INTERFACE -mavx2 -fPIC -DFMT_HEADER_ONLY -fvisibility-inlines-hidden -fno-lto) -if(TT_METAL_VERSIM_DISABLED) - target_compile_options(compiler_flags INTERFACE -DTT_METAL_VERSIM_DISABLED) -endif() if(ENABLE_CODE_TIMERS) target_compile_options(compiler_flags INTERFACE -DTT_ENABLE_CODE_TIMERS) endif() @@ -160,7 +163,7 @@ endif() add_library(metal_header_directories INTERFACE) target_include_directories(metal_header_directories INTERFACE ${PROJECT_SOURCE_DIR}/tt_metal/hw/inc) -target_include_directories(metal_header_directories SYSTEM INTERFACE ${reflect_SOURCE_DIR}) +target_include_directories(metal_header_directories SYSTEM INTERFACE ${reflect_SOURCE_DIR} ${flatbuffers_SOURCE_DIR}/include ${nanomsg_SOURCE_DIR}/include) foreach(lib ${BoostPackages}) target_include_directories(metal_header_directories INTERFACE ${Boost${lib}_SOURCE_DIR}/include) endforeach() diff --git a/cmake/dependencies.cmake b/cmake/dependencies.cmake index b226bd6a9d2..945f937f75f 100644 --- a/cmake/dependencies.cmake +++ b/cmake/dependencies.cmake @@ -56,3 +56,38 @@ CPMAddPackage( GITHUB_REPOSITORY boost-ext/reflect GIT_TAG v1.1.1 ) + +############################################################################################################################ +# Packages needed for tt-metal simulator +# NNG for IPC/TCP communication +# Google Flatbuffers for serialization +# libuv for process mgmt +############################################################################################################################ +CPMAddPackage( + NAME nanomsg + GITHUB_REPOSITORY nanomsg/nng + GIT_TAG v1.8.0 + OPTIONS + "BUILD_SHARED_LIBS ON" + "NNG_TESTS OFF" + "NNG_TOOLS OFF" +) +CPMAddPackage( + NAME flatbuffers + GITHUB_REPOSITORY google/flatbuffers + GIT_TAG v24.3.25 + OPTIONS + "FLATBUFFERS_BUILD_FLATC OFF" + "FLATBUFFERS_BUILD_TESTS OFF" + "FLATBUFFERS_INSTALL OFF" + "FLATBUFFERS_BUILD_FLATLIB OFF" + "FLATBUFFERS_SKIP_MONSTER_EXTRA ON" + "FLATBUFFERS_STRICT_MODE ON" +) +CPMAddPackage( + NAME libuv + GITHUB_REPOSITORY libuv/libuv + GIT_TAG v1.48.0 + OPTIONS + "LIBUV_BUILD_TESTS OFF" +) diff --git a/cmake/umd_device.cmake b/cmake/umd_device.cmake index 2d3b491dc92..ee3e7d54aed 100644 --- a/cmake/umd_device.cmake +++ b/cmake/umd_device.cmake @@ -1,3 +1,21 @@ + +add_library(simulation STATIC + ${UMD_HOME}/device/simulation/tt_simulation_device.cpp + ${UMD_HOME}/device/simulation/tt_simulation_host.cpp +) +target_link_libraries(simulation + PUBLIC compiler_flags + PRIVATE nng uv +) +target_include_directories(simulation PUBLIC + ${UMD_HOME} + ${UMD_HOME}/device + ${UMD_HOME}/third_party/fmt/include + ${flatbuffers_SOURCE_DIR}/include + ${nanomsg_SOURCE_DIR}/include + ${libuv_SOURCE_DIR}/include +) + set(UMD_SRC ${UMD_HOME}/device/architecture_implementation.cpp ${UMD_HOME}/device/blackhole_implementation.cpp @@ -10,10 +28,12 @@ set(UMD_SRC ${UMD_HOME}/device/tt_silicon_driver.cpp ${UMD_HOME}/device/tt_silicon_driver_common.cpp ${UMD_HOME}/device/tt_soc_descriptor.cpp - ${UMD_HOME}/device/tt_versim_stub.cpp ${UMD_HOME}/device/wormhole_implementation.cpp ) add_library(umd_device STATIC ${UMD_SRC}) -target_include_directories(umd_device PRIVATE ${UMD_HOME} ${UMD_HOME}/third_party/fmt/include) -target_link_libraries(umd_device PRIVATE yaml-cpp::yaml-cpp Boost::interprocess rt compiler_flags) +target_include_directories(umd_device PRIVATE + ${UMD_HOME} + ${UMD_HOME}/device + ${UMD_HOME}/third_party/fmt/include) +target_link_libraries(umd_device PRIVATE yaml-cpp::yaml-cpp Boost::interprocess rt simulation) diff --git a/tests/tt_metal/tt_metal/perf_microbenchmark/common/util.hpp b/tests/tt_metal/tt_metal/perf_microbenchmark/common/util.hpp index c625ad841c3..8bf5452d5ce 100644 --- a/tests/tt_metal/tt_metal/perf_microbenchmark/common/util.hpp +++ b/tests/tt_metal/tt_metal/perf_microbenchmark/common/util.hpp @@ -66,11 +66,7 @@ inline uint64_t get_t0_to_any_riscfw_end_cycle(tt::tt_metal::Device *device, con } inline int get_tt_npu_clock(tt::tt_metal::Device *device) { - int ai_clk = 0; -#ifdef TT_METAL_VERSIM_DISABLED - ai_clk = tt::Cluster::instance().get_device_aiclk(device->id()); -#endif - return ai_clk; + return tt::Cluster::instance().get_device_aiclk(device->id()); } template diff --git a/tt_metal/common/base.hpp b/tt_metal/common/base.hpp index dc08af7964d..7000e13cd9d 100644 --- a/tt_metal/common/base.hpp +++ b/tt_metal/common/base.hpp @@ -42,10 +42,8 @@ namespace tt */ enum class TargetDevice : uint8_t { - Model = 0, - Versim = 1, - Silicon = 2, - Golden = 3, + Silicon = 0, + Simulator = 1, Invalid = 0xFF, }; diff --git a/tt_metal/common/core_descriptor.hpp b/tt_metal/common/core_descriptor.hpp index c3e85dd8419..aba6e5ea213 100644 --- a/tt_metal/common/core_descriptor.hpp +++ b/tt_metal/common/core_descriptor.hpp @@ -35,19 +35,16 @@ inline std::string get_core_descriptor_file(const tt::ARCH &arch) { wh_arch = "tt_metal/core_descriptors/"; wh_arch += getenv("WH_ARCH_YAML"); } - bool targeting_versim = false; -#ifndef TT_METAL_VERSIM_DISABLED - targeting_versim = true; -#endif - - if (targeting_versim) { + + bool targeting_sim = std::getenv("TT_METAL_SIMULATOR_EN") != nullptr; + if (targeting_sim) { switch (arch) { case tt::ARCH::Invalid: throw std::runtime_error("Invalid arch not supported"); // will be overwritten in tt_global_state constructor case tt::ARCH::JAWBRIDGE: throw std::runtime_error("JAWBRIDGE arch not supported"); - case tt::ARCH::GRAYSKULL: return tt_metal_home + "tt_metal/core_descriptors/grayskull_versim_1x1_arch.yaml"; - case tt::ARCH::WORMHOLE: throw std::runtime_error("WORMHOLE arch not supported"); - case tt::ARCH::WORMHOLE_B0: return tt_metal_home + "tt_metal/core_descriptors/wormhole_b0_versim_1x1_arch.yaml"; - case tt::ARCH::BLACKHOLE: return tt_metal_home + "tt_metal/core_descriptors/blackhole_versim_1x1_arch.yaml"; + case tt::ARCH::GRAYSKULL: throw std::runtime_error("GRAYSKULL arch not supported for simulator"); + case tt::ARCH::WORMHOLE: throw std::runtime_error("WORMHOLE arch not supported for simulator"); + case tt::ARCH::WORMHOLE_B0: throw std::runtime_error("WORMHOLE_B0 arch not supported for simulator"); + case tt::ARCH::BLACKHOLE: return tt_metal_home + "tt_metal/core_descriptors/blackhole_simulation_1x2_arch.yaml"; default: throw std::runtime_error("Unsupported device arch"); }; } else { diff --git a/tt_metal/common/test_common.hpp b/tt_metal/common/test_common.hpp index b738460f425..6f97b7bb90d 100644 --- a/tt_metal/common/test_common.hpp +++ b/tt_metal/common/test_common.hpp @@ -23,7 +23,6 @@ inline std::string get_soc_description_file(const tt::ARCH &arch, tt::TargetDevice target_device, string output_dir = "") { // Ability to skip this runtime opt, since trimmed SOC desc limits which DRAM channels are available. - bool use_full_soc_desc = getenv("TT_METAL_VERSIM_FORCE_FULL_SOC_DESC"); string tt_metal_home; if (getenv("TT_METAL_HOME")) { tt_metal_home = getenv("TT_METAL_HOME"); @@ -33,14 +32,14 @@ inline std::string get_soc_description_file(const tt::ARCH &arch, tt::TargetDevi if (tt_metal_home.back() != '/') { tt_metal_home += "/"; } - if (target_device == tt::TargetDevice::Versim && !use_full_soc_desc) { + if (target_device == tt::TargetDevice::Simulator){ switch (arch) { - case tt::ARCH::Invalid: throw std::runtime_error("Invalid arch not supported"); // will be overwritten in tt_global_state constructor + case tt::ARCH::Invalid: throw std::runtime_error("Invalid arch not supported"); case tt::ARCH::JAWBRIDGE: throw std::runtime_error("JAWBRIDGE arch not supported"); - case tt::ARCH::GRAYSKULL: return tt_metal_home + "tt_metal/soc_descriptors/grayskull_versim_1x1_arch.yaml"; + case tt::ARCH::GRAYSKULL: throw std::runtime_error("GRAYSKULL arch not supported"); case tt::ARCH::WORMHOLE: throw std::runtime_error("WORMHOLE arch not supported"); - case tt::ARCH::WORMHOLE_B0: return tt_metal_home + "tt_metal/soc_descriptors/wormhole_b0_versim_1x1_arch.yaml"; - case tt::ARCH::BLACKHOLE: return tt_metal_home + "tt_metal/soc_descriptors/blackhole_versim_1x1_arch.yaml"; + case tt::ARCH::WORMHOLE_B0: throw std::runtime_error("WORMHOLE_B0 arch not supported"); + case tt::ARCH::BLACKHOLE: return tt_metal_home + "tt_metal/soc_descriptors/blackhole_simulation_1x2_arch.yaml"; default: throw std::runtime_error("Unsupported device arch"); }; } else { diff --git a/tt_metal/core_descriptors/blackhole_simulation_1x2_arch.yaml b/tt_metal/core_descriptors/blackhole_simulation_1x2_arch.yaml new file mode 100644 index 00000000000..1234ec50bb6 --- /dev/null +++ b/tt_metal/core_descriptors/blackhole_simulation_1x2_arch.yaml @@ -0,0 +1,34 @@ +# Anything using [[#, #]] is logical coordinates (Can be relative) +# relative index: 0 means first row, -1 means last row of functional grid... + +# product name: +# num of HW command queues: +# core descriptor config + +blackhole: + 1: + compute_with_storage_grid_range: # Logical only start and end [x, y] + start: [0, 1] + end: [1, 1] + + storage_cores: # Relative to grid of tensix cores + [] + + dispatch_cores: + [] + + dispatch_core_type: + "tensix" + 2: + compute_with_storage_grid_range: # Logical only start and end [x, y] + start: [0, 1] + end: [1, 1] + + storage_cores: # Relative to grid of tensix cores + [] + + dispatch_cores: + [] + + dispatch_core_type: + "tensix" diff --git a/tt_metal/hw/inc/blackhole/dev_mem_map_versim.h b/tt_metal/hw/inc/blackhole/dev_mem_map_versim.h deleted file mode 100644 index a3243bad922..00000000000 --- a/tt_metal/hw/inc/blackhole/dev_mem_map_versim.h +++ /dev/null @@ -1,11 +0,0 @@ -// SPDX-FileCopyrightText: © 2023 Tenstorrent Inc. -// -// SPDX-License-Identifier: Apache-2.0 - -#pragma once - -#include "dev_mem_map.h" -// This is to support some deprecated mappings for versim to get build up -#ifndef TT_METAL_VERSIM_DISABLED -#define TEST_MAILBOX_ADDRESS MEM_MAILBOX_BASE -#endif diff --git a/tt_metal/hw/inc/grayskull/dev_mem_map_versim.h b/tt_metal/hw/inc/grayskull/dev_mem_map_versim.h deleted file mode 100644 index a3243bad922..00000000000 --- a/tt_metal/hw/inc/grayskull/dev_mem_map_versim.h +++ /dev/null @@ -1,11 +0,0 @@ -// SPDX-FileCopyrightText: © 2023 Tenstorrent Inc. -// -// SPDX-License-Identifier: Apache-2.0 - -#pragma once - -#include "dev_mem_map.h" -// This is to support some deprecated mappings for versim to get build up -#ifndef TT_METAL_VERSIM_DISABLED -#define TEST_MAILBOX_ADDRESS MEM_MAILBOX_BASE -#endif diff --git a/tt_metal/hw/inc/grayskull/tensix.h b/tt_metal/hw/inc/grayskull/tensix.h index c7ea1c3f688..4774048fe8f 100644 --- a/tt_metal/hw/inc/grayskull/tensix.h +++ b/tt_metal/hw/inc/grayskull/tensix.h @@ -11,7 +11,6 @@ #include #include "cfg_defines.h" #include "dev_mem_map.h" -#include "dev_mem_map_versim.h" // Convenience and type defines typedef std::uint32_t uint; diff --git a/tt_metal/hw/inc/wormhole/dev_mem_map_versim.h b/tt_metal/hw/inc/wormhole/dev_mem_map_versim.h deleted file mode 100644 index a3243bad922..00000000000 --- a/tt_metal/hw/inc/wormhole/dev_mem_map_versim.h +++ /dev/null @@ -1,11 +0,0 @@ -// SPDX-FileCopyrightText: © 2023 Tenstorrent Inc. -// -// SPDX-License-Identifier: Apache-2.0 - -#pragma once - -#include "dev_mem_map.h" -// This is to support some deprecated mappings for versim to get build up -#ifndef TT_METAL_VERSIM_DISABLED -#define TEST_MAILBOX_ADDRESS MEM_MAILBOX_BASE -#endif diff --git a/tt_metal/hw/inc/wormhole/tensix.h b/tt_metal/hw/inc/wormhole/tensix.h index 92da2b85b9b..83d781e263c 100644 --- a/tt_metal/hw/inc/wormhole/tensix.h +++ b/tt_metal/hw/inc/wormhole/tensix.h @@ -11,7 +11,6 @@ #include #include "cfg_defines.h" #include "dev_mem_map.h" -#include "dev_mem_map_versim.h" // Convenience and type defines typedef std::uint32_t uint; diff --git a/tt_metal/impl/device/device.cpp b/tt_metal/impl/device/device.cpp index 5785ec4cfdb..e0635d2b60b 100644 --- a/tt_metal/impl/device/device.cpp +++ b/tt_metal/impl/device/device.cpp @@ -147,10 +147,8 @@ void Device::initialize_cluster() { if (llrt::OptionsG.get_clear_l1()) { this->clear_l1_state(); } -#ifdef TT_METAL_VERSIM_DISABLED int ai_clk = tt::Cluster::instance().get_device_aiclk(this->id_); log_info(tt::LogMetal, "AI CLK for device {} is: {} MHz", this->id_, ai_clk); -#endif } void Device::initialize_allocator(size_t l1_small_size, size_t trace_region_size, const std::vector &l1_bank_remap) { diff --git a/tt_metal/llrt/llrt.hpp b/tt_metal/llrt/llrt.hpp index f664765ad71..a87768dea7c 100644 --- a/tt_metal/llrt/llrt.hpp +++ b/tt_metal/llrt/llrt.hpp @@ -13,7 +13,7 @@ // clang-format off #include "llrt/tt_cluster.hpp" #include "tensix.h" -#include "tt_metal/third_party/umd/device/device_api.h" +#include "tt_metal/third_party/umd/device/device_api_metal.h" #include "tt_metal/third_party/umd/device/tt_xy_pair.h" #include "llrt_common/tiles.hpp" #include "llrt/tt_memory.h" diff --git a/tt_metal/llrt/tlb_config.hpp b/tt_metal/llrt/tlb_config.hpp index e0aae5a23a1..099b530a04b 100644 --- a/tt_metal/llrt/tlb_config.hpp +++ b/tt_metal/llrt/tlb_config.hpp @@ -4,7 +4,7 @@ #pragma once -#include "third_party/umd/device/device_api.h" +#include "third_party/umd/device/device_api_metal.h" #include "tt_metal/common/tt_backend_api_types.hpp" #include "tt_metal/common/metal_soc_descriptor.h" diff --git a/tt_metal/llrt/tt_cluster.cpp b/tt_metal/llrt/tt_cluster.cpp index 77d29db2ec7..f31ec38325d 100644 --- a/tt_metal/llrt/tt_cluster.cpp +++ b/tt_metal/llrt/tt_cluster.cpp @@ -49,27 +49,26 @@ Cluster::Cluster() { } void Cluster::detect_arch_and_target() { -#ifdef TT_METAL_VERSIM_DISABLED - this->target_type_ = TargetDevice::Silicon; - std::vector physical_mmio_device_ids = tt_SiliconDevice::detect_available_device_ids(); - this->arch_ = detect_arch(physical_mmio_device_ids.at(0)); - for (int dev_index = 1; dev_index < physical_mmio_device_ids.size(); dev_index++) { - chip_id_t device_id = physical_mmio_device_ids.at(dev_index); - tt::ARCH detected_arch = detect_arch(device_id); - TT_FATAL( - this->arch_ == detected_arch, - "Expected all devices to be {} but device {} is {}", - get_arch_str(this->arch_), - device_id, - get_arch_str(detected_arch)); + if(std::getenv("TT_METAL_SIMULATOR_EN")) { + this->target_type_ = TargetDevice::Simulator; + auto arch_env = getenv("ARCH_NAME"); + TT_FATAL(arch_env, "ARCH_NAME env var needed for VCS"); + this->arch_ = tt::get_arch_from_string(arch_env); + }else { + this->target_type_ = TargetDevice::Silicon; + std::vector physical_mmio_device_ids = tt_SiliconDevice::detect_available_device_ids(); + this->arch_ = detect_arch(physical_mmio_device_ids.at(0)); + for (int dev_index = 1; dev_index < physical_mmio_device_ids.size(); dev_index++) { + chip_id_t device_id = physical_mmio_device_ids.at(dev_index); + tt::ARCH detected_arch = detect_arch(device_id); + TT_FATAL( + this->arch_ == detected_arch, + "Expected all devices to be {} but device {} is {}", + get_arch_str(this->arch_), + device_id, + get_arch_str(detected_arch)); + } } -#else - this->target_type_ = TargetDevice::Versim; - auto arch_env = getenv("ARCH_NAME"); - TT_FATAL(arch_env, "arch_env needs to be set for versim (ARCH_NAME=)"); - this->arch_ = tt::get_arch_from_string(arch_env); -#endif - #ifdef ARCH_GRAYSKULL TT_FATAL( this->arch_ == tt::ARCH::GRAYSKULL, @@ -89,7 +88,7 @@ void Cluster::detect_arch_and_target() { get_string(this->arch_)); #endif - TT_FATAL(this->target_type_ == TargetDevice::Versim or this->target_type_ == TargetDevice::Silicon); + TT_FATAL(this->target_type_ == TargetDevice::Silicon or this->target_type_ == TargetDevice::Simulator); } std::filesystem::path get_cluster_desc_yaml() { @@ -140,8 +139,13 @@ void Cluster::generate_cluster_descriptor() { // create-eth-map not available for Blackhole bring up if (this->arch_ == tt::ARCH::GRAYSKULL or this->arch_ == tt::ARCH::BLACKHOLE) { // Cannot use tt_SiliconDevice::detect_available_device_ids because that returns physical device IDs - std::vector physical_mmio_device_ids = tt_SiliconDevice::detect_available_device_ids(); + std::vector physical_mmio_device_ids; std::set logical_mmio_device_ids; + if (this->target_type_ == TargetDevice::Simulator) { + physical_mmio_device_ids = tt_SimulationDevice::detect_available_device_ids(); + } else{ + physical_mmio_device_ids = tt_SiliconDevice::detect_available_device_ids(); + } for (chip_id_t logical_mmio_device_id = 0; logical_mmio_device_id < physical_mmio_device_ids.size(); logical_mmio_device_id++) { logical_mmio_device_ids.insert(logical_mmio_device_id); @@ -159,9 +163,9 @@ void Cluster::generate_cluster_descriptor() { } // Use cluster descriptor to map MMIO device id to all devices on the same card (including the MMIO device) - if (this->target_type_ == TargetDevice::Versim) { - std::set dummy_versim_card = {0}; - this->devices_grouped_by_assoc_mmio_device_[0] = dummy_versim_card; + if (this->target_type_ == TargetDevice::Simulator) { + std::set dummy_card = {0}; + this->devices_grouped_by_assoc_mmio_device_[0] = dummy_card; this->device_to_mmio_device_[0] = 0; } else { for (chip_id_t device_id : this->cluster_desc_->get_all_chips()) { @@ -200,10 +204,6 @@ void Cluster::initialize_device_drivers() { this->open_driver(mmio_device_id, controlled_devices); tt_device_params default_params; - if (getenv("TT_METAL_VERSIM_DUMP_CORES")) { - std::string dump_cores_string = getenv("TT_METAL_VERSIM_DUMP_CORES"); - default_params.vcd_dump_cores = tt::utils::strsplit(dump_cores_string, ','); - } this->start_driver(mmio_device_id, default_params); } } @@ -276,8 +276,8 @@ void Cluster::open_driver( // Adding this check is a workaround for current UMD bug that only uses this getter to populate private metadata // that is later expected to be populated by unrelated APIs TT_FATAL(device_driver->get_target_mmio_device_ids().size() == 1); - } else if (this->target_type_ == TargetDevice::Versim) { - device_driver = std::make_unique(sdesc_path, this->cluster_desc_path_); + } else if (this->target_type_ == TargetDevice::Simulator) { + device_driver = std::make_unique(sdesc_path); } device_driver->set_device_dram_address_params(dram_address_params); device_driver->set_device_l1_address_params(l1_address_params); @@ -302,7 +302,6 @@ void Cluster::start_driver(chip_id_t mmio_device_id, tt_device_params &device_pa Cluster::~Cluster() { log_info(tt::LogDevice, "Closing user mode device drivers"); - for (const auto &[mmio_device_id, device_driver] : this->mmio_device_id_to_driver_) { device_driver->close_device(); } @@ -333,7 +332,7 @@ const metal_SocDescriptor &Cluster::get_soc_desc(chip_id_t chip) const { } uint32_t Cluster::get_harvested_rows(chip_id_t chip) const { - if (this->target_type_ == TargetDevice::Versim) { + if (this->target_type_ == TargetDevice::Simulator) { return 0; } else { return this->get_driver(chip).harvested_rows_per_target.at(chip); diff --git a/tt_metal/llrt/tt_cluster.hpp b/tt_metal/llrt/tt_cluster.hpp index 1ddd625e360..05b4cca0344 100644 --- a/tt_metal/llrt/tt_cluster.hpp +++ b/tt_metal/llrt/tt_cluster.hpp @@ -13,7 +13,7 @@ #include "common/tt_backend_api_types.hpp" #include "host_mem_address_map.h" #include "hostdevcommon/common_runtime_address_map.h" -#include "third_party/umd/device/device_api.h" +#include "third_party/umd/device/device_api_metal.h" #include "tt_metal/third_party/umd/device/tt_cluster_descriptor.h" #include "tt_metal/third_party/umd/device/tt_xy_pair.h" diff --git a/tt_metal/soc_descriptors/grayskull_versim_1x1_arch.yaml b/tt_metal/soc_descriptors/blackhole_simulation_1x2_arch.yaml similarity index 64% rename from tt_metal/soc_descriptors/grayskull_versim_1x1_arch.yaml rename to tt_metal/soc_descriptors/blackhole_simulation_1x2_arch.yaml index 4b4ab3fa272..8230b041df6 100644 --- a/tt_metal/soc_descriptors/grayskull_versim_1x1_arch.yaml +++ b/tt_metal/soc_descriptors/blackhole_simulation_1x2_arch.yaml @@ -1,8 +1,3 @@ -# soc-descriptor yaml -# Anything using [#-#] is noc coordinates -# Anything using [[#, #]] is logical coordinates (Can be relative) -# relative index: 0 means first row, -1 means last row of functional grid... - grid: x_size: 2 y_size: 2 @@ -29,20 +24,16 @@ eth: [] functional_workers: - [ - 1-1 - ] - + [0-1, 1-1] + harvested_workers: [] router_only: - [ - 0-0, 0-1 - ] + [0-0] worker_l1_size: - 1048576 + 1499136 dram_bank_size: 1073741824 @@ -50,7 +41,7 @@ dram_bank_size: eth_l1_size: 0 -arch_name: GRAYSKULL +arch_name: BLACKHOLE features: unpacker: diff --git a/tt_metal/soc_descriptors/blackhole_versim_1x1_arch.yaml b/tt_metal/soc_descriptors/blackhole_versim_1x1_arch.yaml deleted file mode 100644 index aca55c5fc49..00000000000 --- a/tt_metal/soc_descriptors/blackhole_versim_1x1_arch.yaml +++ /dev/null @@ -1,67 +0,0 @@ -# soc-descriptor yaml -# Anything using [#-#] is noc coordinates -# Anything using [[#, #]] is logical coordinates (Can be relative) -# relative index: 0 means first row, -1 means last row of functional grid... -grid: - x_size: 17 - y_size: 12 - -arc: - [ 8-0 ] - -pcie: - [ 11-0 ] - -dram: - [ - [0-0, 0-1], - ] - -dram_preferred_eth_endpoint: - [ 0-0 ] - -dram_preferred_worker_endpoint: - [ 0-1 ] - -dram_address_offsets: - [ 0 ] - -eth: - [ ] - -functional_workers: - [ - 1-2 - ] - -harvested_workers: - [] - -router_only: - [ - 1-0 - ] - -worker_l1_size: - 1499136 - -dram_bank_size: - 4294967296 - -eth_l1_size: - 262144 - -arch_name: BLACKHOLE - -features: - noc: - translation_id_enabled: True - unpacker: - version: 2 - inline_srca_trans_without_srca_trans_instr: True - math: - dst_size_alignment: 32768 - packer: - version: 2 - overlay: - version: 2 diff --git a/tt_metal/soc_descriptors/wormhole_b0_versim_1x1_arch.yaml b/tt_metal/soc_descriptors/wormhole_b0_versim_1x1_arch.yaml deleted file mode 100644 index be4cdc51904..00000000000 --- a/tt_metal/soc_descriptors/wormhole_b0_versim_1x1_arch.yaml +++ /dev/null @@ -1,65 +0,0 @@ -# soc-descriptor yaml -# Anything using [#-#] is noc coordinates -# Anything using [[#, #]] is logical coordinates (Can be relative) -# relative index: 0 means first row, -1 means last row of functional grid... -grid: - x_size: 2 - y_size: 2 - -arc: - [ ] - -pcie: - [ ] - -dram: - [ - [0-0, 0-1], - ] - -dram_preferred_eth_endpoint: - [ 0-0 ] - -dram_preferred_worker_endpoint: - [ 0-1 ] - -dram_address_offsets: - [ 0, 1073741824, 0, 1073741824, 0, 1073741824, 0, 1073741824, 0, 1073741824, 0, 1073741824 ] - -eth: - [ ] - -functional_workers: - [ - 1-1, - ] - -harvested_workers: - [] - -router_only: - [ - 1-0 - ] - -worker_l1_size: - 1499136 - -dram_bank_size: - 1073741824 - -eth_l1_size: - 262144 - -arch_name: WORMHOLE_B0 - -features: - unpacker: - version: 2 - inline_srca_trans_without_srca_trans_instr: True - math: - dst_size_alignment: 32768 - packer: - version: 2 - overlay: - version: 2 diff --git a/tt_metal/third_party/umd b/tt_metal/third_party/umd index 80cd1705c5a..ed64cd93b51 160000 --- a/tt_metal/third_party/umd +++ b/tt_metal/third_party/umd @@ -1 +1 @@ -Subproject commit 80cd1705c5a9d259ce87a6c9949aea1080760236 +Subproject commit ed64cd93b5191afd2a43419a75166b4300f5574f diff --git a/tt_metal/tt_metal.cpp b/tt_metal/tt_metal.cpp index b4bbec6146b..632af0aee2b 100644 --- a/tt_metal/tt_metal.cpp +++ b/tt_metal/tt_metal.cpp @@ -700,19 +700,11 @@ void EnableAllocs(Device *device) { tt::tt_metal::allocator::enable_allocs(*(dev } // namespace detail size_t GetNumAvailableDevices() { -#ifdef TT_METAL_VERSIM_DISABLED return tt::Cluster::instance().number_of_user_devices(); -#else - return 1; -#endif } size_t GetNumPCIeDevices() { -#ifdef TT_METAL_VERSIM_DISABLED return tt::Cluster::instance().number_of_pci_devices(); -#else - return 1; -#endif } chip_id_t GetPCIeDeviceID(chip_id_t device_id){ diff --git a/ttnn/CMakeLists.txt b/ttnn/CMakeLists.txt index f9714e1aaa1..9d64d243efe 100644 --- a/ttnn/CMakeLists.txt +++ b/ttnn/CMakeLists.txt @@ -167,7 +167,7 @@ set(TTNN_PRECOMPILED_HEADERS ${PROJECT_SOURCE_DIR}/tt_metal/third_party/tracy/public/tracy/Tracy.hpp ${PROJECT_SOURCE_DIR}/tt_metal/third_party/fmt/fmt/core.h ${PROJECT_SOURCE_DIR}/tt_metal/third_party/fmt/fmt/format.h - ${PROJECT_SOURCE_DIR}/tt_metal/third_party/umd/device/device_api.h + ${PROJECT_SOURCE_DIR}/tt_metal/third_party/umd/device/device_api_metal.h ${PROJECT_SOURCE_DIR}/tt_metal/third_party/umd/device/tt_device.h