From 5381e4370ab2b67e8dd8a1fa3f938b222930a258 Mon Sep 17 00:00:00 2001 From: =?UTF-8?q?Bojan=20Ro=C5=A1ko?= <156314064+broskoTT@users.noreply.github.com> Date: Tue, 17 Dec 2024 13:08:13 +0100 Subject: [PATCH] [UMD] Removed set_*_params calls and constants (#15908) ### Ticket Related to https://github.com/tenstorrent/tt-metal/issues/13948 ### Problem description Related to UMD change: https://github.com/tenstorrent/tt-umd/pull/395 These parameters are defined by hardware. The only parameters to be defined by the client are membars. This change adds set_barrier_address_params accordingly. ### What's changed Removed set_device_dram_address_params and set_device_l1_address_params, and related constants. Replace with call to set_barrier_address_params. ### Checklist - [ ] All post-commit tests : https://github.com/tenstorrent/tt-metal/actions/runs/12361649190 - main generally unhealthy, I haven't seen new fails related to this change. - [x] Blackhole post-commit tests : https://github.com/tenstorrent/tt-metal/actions/runs/12361650912 - [ ] (Single-card) Model perf tests : https://github.com/tenstorrent/tt-metal/actions/runs/12361652594 - N300 job failing on main - [x] (Single-card) Device perf regressions : https://github.com/tenstorrent/tt-metal/actions/runs/12361654455 - [x] (T3K) T3000 unit tests : https://github.com/tenstorrent/tt-metal/actions/runs/12361656223 - [ ] (T3K) T3000 demo tests : https://github.com/tenstorrent/tt-metal/actions/runs/12361657485 - falcon job failing on main - [ ] (TG) TG unit tests : https://github.com/tenstorrent/tt-metal/actions/runs/12361658889 - unit test failing on main - [ ] (TG) TG demo tests : https://github.com/tenstorrent/tt-metal/actions/runs/12361660547 - failing on main - [ ] (TGG) TGG unit tests : https://github.com/tenstorrent/tt-metal/actions/runs/12361662343 - failing on main for a long time due to hugepages - [x] (TGG) TGG demo tests : https://github.com/tenstorrent/tt-metal/actions/runs/12361664160 --- tt_metal/llrt/blackhole/bh_hal_active_eth.cpp | 3 --- tt_metal/llrt/hal.hpp | 3 +-- tt_metal/llrt/tt_cluster.cpp | 12 ++++++------ tt_metal/llrt/tt_cluster.hpp | 2 -- tt_metal/llrt/wormhole/wh_hal_active_eth.cpp | 3 --- tt_metal/third_party/umd | 2 +- 6 files changed, 8 insertions(+), 17 deletions(-) diff --git a/tt_metal/llrt/blackhole/bh_hal_active_eth.cpp b/tt_metal/llrt/blackhole/bh_hal_active_eth.cpp index f6620851b27..bf86f596628 100644 --- a/tt_metal/llrt/blackhole/bh_hal_active_eth.cpp +++ b/tt_metal/llrt/blackhole/bh_hal_active_eth.cpp @@ -44,8 +44,6 @@ HalCoreInfoType create_active_eth_mem_map() { mem_map_bases[static_cast(HalL1MemAddrType::GO_MSG)] = GET_ETH_MAILBOX_ADDRESS_HOST(go_message); mem_map_bases[static_cast(HalL1MemAddrType::LAUNCH_MSG_BUFFER_RD_PTR)] = GET_ETH_MAILBOX_ADDRESS_HOST(launch_msg_rd_ptr); - mem_map_bases[static_cast(HalL1MemAddrType::FW_VERSION_ADDR)] = - eth_l1_mem::address_map::FW_VERSION_ADDR; mem_map_bases[static_cast(HalL1MemAddrType::BANK_TO_NOC_SCRATCH)] = eth_l1_mem::address_map::ERISC_MEM_BANK_TO_NOC_SCRATCH; @@ -66,7 +64,6 @@ HalCoreInfoType create_active_eth_mem_map() { eth_l1_mem::address_map::MAX_SIZE - eth_l1_mem::address_map::ERISC_L1_UNRESERVED_BASE; mem_map_sizes[static_cast(HalL1MemAddrType::GO_MSG)] = sizeof(go_msg_t); mem_map_sizes[static_cast(HalL1MemAddrType::LAUNCH_MSG_BUFFER_RD_PTR)] = sizeof(std::uint32_t); - mem_map_sizes[static_cast(HalL1MemAddrType::FW_VERSION_ADDR)] = sizeof(std::uint32_t); mem_map_sizes[static_cast(HalL1MemAddrType::BANK_TO_NOC_SCRATCH)] = eth_l1_mem::address_map::ERISC_MEM_BANK_TO_NOC_SIZE; diff --git a/tt_metal/llrt/hal.hpp b/tt_metal/llrt/hal.hpp index 9344b6bd4ac..bc23f59e270 100644 --- a/tt_metal/llrt/hal.hpp +++ b/tt_metal/llrt/hal.hpp @@ -49,10 +49,9 @@ enum class HalL1MemAddrType : uint8_t { CORE_INFO, GO_MSG, LAUNCH_MSG_BUFFER_RD_PTR, - FW_VERSION_ADDR, // Really only applicable to active eth core right now LOCAL, BANK_TO_NOC_SCRATCH, - COUNT // Keep this last so it always indicates number of enum options + COUNT // Keep this last so it always indicates number of enum options }; enum class HalDramMemAddrType : uint8_t { DRAM_BARRIER = 0, COUNT = 1 }; diff --git a/tt_metal/llrt/tt_cluster.cpp b/tt_metal/llrt/tt_cluster.cpp index bf056fd446a..12f144dd7ce 100644 --- a/tt_metal/llrt/tt_cluster.cpp +++ b/tt_metal/llrt/tt_cluster.cpp @@ -252,15 +252,15 @@ void Cluster::open_driver(const bool &skip_driver_allocs) { } else if (this->target_type_ == TargetDevice::Simulator) { device_driver = std::make_unique(sdesc_path); } - std::uint32_t dram_barrier_base = tt_metal::hal.get_dev_addr(tt_metal::HalDramMemAddrType::DRAM_BARRIER); - device_driver->set_device_dram_address_params(tt_device_dram_address_params{dram_barrier_base}); - l1_address_params.tensix_l1_barrier_base = tt_metal::hal.get_dev_addr(tt_metal::HalProgrammableCoreType::TENSIX, tt_metal::HalL1MemAddrType::BARRIER); + barrier_address_params barrier_params; + barrier_params.tensix_l1_barrier_base = tt_metal::hal.get_dev_addr(tt_metal::HalProgrammableCoreType::TENSIX, tt_metal::HalL1MemAddrType::BARRIER); + barrier_params.dram_barrier_base = tt_metal::hal.get_dev_addr(tt_metal::HalDramMemAddrType::DRAM_BARRIER); + if (tt_metal::hal.get_arch() != tt::ARCH::GRAYSKULL) { - l1_address_params.eth_l1_barrier_base = tt_metal::hal.get_dev_addr(tt_metal::HalProgrammableCoreType::ACTIVE_ETH, tt_metal::HalL1MemAddrType::BARRIER); - l1_address_params.fw_version_addr = tt_metal::hal.get_dev_addr(tt_metal::HalProgrammableCoreType::ACTIVE_ETH, tt_metal::HalL1MemAddrType::FW_VERSION_ADDR); + barrier_params.eth_l1_barrier_base = tt_metal::hal.get_dev_addr(tt_metal::HalProgrammableCoreType::ACTIVE_ETH, tt_metal::HalL1MemAddrType::BARRIER); } - device_driver->set_device_l1_address_params(l1_address_params); + device_driver->set_barrier_address_params(barrier_params); this->get_metal_desc_from_tt_desc( device_driver->get_virtual_soc_descriptors(), device_driver->get_harvesting_masks_for_soc_descriptors()); diff --git a/tt_metal/llrt/tt_cluster.hpp b/tt_metal/llrt/tt_cluster.hpp index 3b59f4cc7d0..629f23ddd23 100644 --- a/tt_metal/llrt/tt_cluster.hpp +++ b/tt_metal/llrt/tt_cluster.hpp @@ -297,8 +297,6 @@ class Cluster { // Mapping of each devices' ethernet routing mode std::unordered_map> device_eth_routing_info_; - tt_device_l1_address_params l1_address_params; - std::unordered_map>> ethernet_sockets_; }; diff --git a/tt_metal/llrt/wormhole/wh_hal_active_eth.cpp b/tt_metal/llrt/wormhole/wh_hal_active_eth.cpp index c0af4cc0bd7..118140f01c1 100644 --- a/tt_metal/llrt/wormhole/wh_hal_active_eth.cpp +++ b/tt_metal/llrt/wormhole/wh_hal_active_eth.cpp @@ -41,8 +41,6 @@ HalCoreInfoType create_active_eth_mem_map() { mem_map_bases[static_cast(HalL1MemAddrType::GO_MSG)] = GET_ETH_MAILBOX_ADDRESS_HOST(go_message); mem_map_bases[static_cast(HalL1MemAddrType::LAUNCH_MSG_BUFFER_RD_PTR)] = GET_ETH_MAILBOX_ADDRESS_HOST(launch_msg_rd_ptr); - mem_map_bases[static_cast(HalL1MemAddrType::FW_VERSION_ADDR)] = - eth_l1_mem::address_map::FW_VERSION_ADDR; mem_map_bases[static_cast(HalL1MemAddrType::BANK_TO_NOC_SCRATCH)] = eth_l1_mem::address_map::ERISC_MEM_BANK_TO_NOC_SCRATCH; @@ -63,7 +61,6 @@ HalCoreInfoType create_active_eth_mem_map() { eth_l1_mem::address_map::MAX_SIZE - eth_l1_mem::address_map::ERISC_L1_UNRESERVED_BASE; mem_map_sizes[static_cast(HalL1MemAddrType::GO_MSG)] = sizeof(go_msg_t); mem_map_sizes[static_cast(HalL1MemAddrType::LAUNCH_MSG_BUFFER_RD_PTR)] = sizeof(uint32_t); - mem_map_sizes[static_cast(HalL1MemAddrType::FW_VERSION_ADDR)] = sizeof(std::uint32_t); mem_map_sizes[static_cast(HalL1MemAddrType::BANK_TO_NOC_SCRATCH)] = eth_l1_mem::address_map::ERISC_MEM_BANK_TO_NOC_SIZE; std::vector> processor_classes(NumEthDispatchClasses); diff --git a/tt_metal/third_party/umd b/tt_metal/third_party/umd index 7da60173887..cfadca1bbba 160000 --- a/tt_metal/third_party/umd +++ b/tt_metal/third_party/umd @@ -1 +1 @@ -Subproject commit 7da601738873f11c0fb946e590e957f5b6c8a8a9 +Subproject commit cfadca1bbbabf2d9cf3ab038322d7416d8059ff3