#20712: Update external cable check to support N300/T3K and add report to health check (#22830)

### Ticket
#20712

### Problem description
Want to report whether missing links are internal/external for health
check.

### What's changed
Add support for checking external cables on generic N300 devices and
update health check script to report which links are internal/external.

### Checklist
- [ ] [All post
commit](https://github.com/tenstorrent/tt-metal/actions/workflows/all-post-commit-workflows.yaml)
CI passes
- [ ] [Blackhole Post
commit](https://github.com/tenstorrent/tt-metal/actions/workflows/blackhole-post-commit.yaml)
CI with demo tests passes (if applicable)
- [ ] [Model
regression](https://github.com/tenstorrent/tt-metal/actions/workflows/perf-models.yaml)
CI passes (if applicable)
- [ ] [Device performance
regression](https://github.com/tenstorrent/tt-metal/actions/workflows/perf-device-models.yaml)
CI passes (if applicable)
- [ ] (For models and ops writers) [Single-card demo
tests](https://github.com/tenstorrent/tt-metal/actions/workflows/single-card-demo-tests.yaml)
CI passes (if applicable) See [recommended dev
flow](https://github.com/tenstorrent/tt-metal/blob/main/models/MODEL_ADD.md#a-recommended-dev-flow-on-github-for-adding-new-models).
- [x] New/Existing tests provide coverage for changes

Co-authored-by: Allan Liu <aliu@tenstorrent.com>

Author: tt-aho

tests/tt_metal/tt_fabric/system_health/test_system_health.cpp

@@ -28,8 +28,7 @@ std::pair<std::uint32_t, std::uint32_t> get_ubb_ids(chip_id_t chip_id) {
     const auto& tray_bus_ids = ubb_bus_ids.at(cluster.arch());
     auto tray_bus_id_it = std::find(tray_bus_ids.begin(), tray_bus_ids.end(), cluster.get_bus_id(chip_id) & 0xF0);
     if (tray_bus_id_it != tray_bus_ids.end()) {
-        auto unique_chip_id = cluster.get_unique_chip_ids().at(chip_id);
-        auto ubb_asic_id = ((unique_chip_id >> 56) & 0xFF);
+        auto ubb_asic_id = cluster.get_ubb_asic_id(chip_id);
         return std::make_pair(tray_bus_id_it - tray_bus_ids.begin() + 1, ubb_asic_id);
     }
     return std::make_pair(0, 0);
@@ -44,7 +43,6 @@ TEST(Cluster, ReportSystemHealth) {
 
     auto unique_chip_ids = tt::tt_metal::MetalContext::instance().get_cluster().get_unique_chip_ids();
     std::stringstream ss;
-    ss << "Found " << unique_chip_ids.size() << " chips in cluster:" << std::endl;
     std::vector<std::uint32_t> read_vec;
     auto retrain_count_addr = tt::tt_metal::MetalContext::instance().hal().get_dev_addr(
         tt::tt_metal::HalProgrammableCoreType::ACTIVE_ETH, tt::tt_metal::HalL1MemAddrType::RETRAIN_COUNT);
@@ -54,6 +52,7 @@ TEST(Cluster, ReportSystemHealth) {
             unique_chip_ids[chip_id] = chip_id;
         }
     }
+    ss << "Found " << unique_chip_ids.size() << " chips in cluster:" << std::endl;
 
     auto cluster_type = cluster.get_cluster_type();
 
@@ -73,27 +72,29 @@ TEST(Cluster, ReportSystemHealth) {
             std::stringstream eth_ss;
             cluster.read_core(read_vec, sizeof(uint32_t), virtual_eth_core, retrain_count_addr);
             eth_ss << " eth channel " << std::dec << (uint32_t)chan << " " << eth_core.str();
+            std::string connection_type =
+                cluster.is_external_cable(chip_id, eth_core) ? "(external connector)" : "(internal trace)";
             if (cluster.is_ethernet_link_up(chip_id, eth_core)) {
                 if (eth_connections.at(chip_id).find(chan) != eth_connections.at(chip_id).end()) {
                     const auto& [connected_chip_id, connected_eth_core] =
                         cluster.get_connected_ethernet_core(std::make_tuple(chip_id, eth_core));
                     std::cout << "Connected chip: " << connected_chip_id
                               << " connected eth core: " << connected_eth_core.str() << std::endl;
-                    eth_ss << " link UP, retrain: " << read_vec[0] << ", connected to chip " << connected_chip_id << " "
-                           << connected_eth_core.str();
+                    eth_ss << " link UP " << connection_type << ", retrain: " << read_vec[0] << ", connected to chip "
+                           << connected_chip_id << " " << connected_eth_core.str();
                 } else {
                     const auto& [connected_chip_unique_id, connected_eth_core] =
                         cluster.get_connected_ethernet_core_to_remote_mmio_device(std::make_tuple(chip_id, eth_core));
                     std::cout << "Connected unique chip: " << connected_chip_unique_id
                               << " connected eth core: " << connected_eth_core.str() << std::endl;
-                    eth_ss << " link UP, retrain: " << read_vec[0] << ", connected to chip " << connected_chip_unique_id
-                           << " " << connected_eth_core.str();
+                    eth_ss << " link UP " << connection_type << ", retrain: " << read_vec[0] << ", connected to chip "
+                           << connected_chip_unique_id << " " << connected_eth_core.str();
                 }
                 if (read_vec[0] > 0) {
                     unexpected_system_states.push_back(chip_id_ss.str() + eth_ss.str());
                 }
             } else {
-                eth_ss << " link DOWN";
+                eth_ss << " link DOWN/unconnected " << connection_type;
                 unexpected_system_states.push_back(chip_id_ss.str() + eth_ss.str());
             }
             ss << eth_ss.str() << std::endl;
@@ -176,9 +177,13 @@ TEST(Cluster, TestMeshFullConnectivity) {
             auto [tray_id, ubb_asic_id] = get_ubb_ids(chip);
             chip_ss << " Tray: " << tray_id << " N" << ubb_asic_id;
         }
+        const auto& soc_desc = cluster.get_soc_desc(chip);
         std::map<chip_id_t, int> num_connections_to_chip;
         for (const auto& [channel, remote_chip_and_channel] : connections) {
-            num_connections_to_chip[std::get<0>(remote_chip_and_channel)]++;
+            tt::umd::CoreCoord logical_active_eth = soc_desc.get_eth_core_for_channel(channel, CoordSystem::LOGICAL);
+            if (cluster.is_ethernet_link_up(chip, logical_active_eth)) {
+                num_connections_to_chip[std::get<0>(remote_chip_and_channel)]++;
+            }
         }
         if (target_system_topology.has_value()) {
             if (*target_system_topology == FabricType::TORUS_2D) {

tt_metal/fabric/control_plane.cpp

@@ -27,12 +27,12 @@
 #include "core_coord.hpp"
 #include "fabric_host_interface.h"
 #include "hal_types.hpp"
+#include "impl/context/metal_context.hpp"
 #include "logger.hpp"
 #include "mesh_coord.hpp"
 #include "mesh_graph.hpp"
 #include "metal_soc_descriptor.h"
 #include "routing_table_generator.hpp"
-#include "impl/context/metal_context.hpp"
 #include <umd/device/tt_core_coordinates.h>
 #include <umd/device/tt_xy_pair.h>
 #include <umd/device/types/cluster_descriptor_types.h>
@@ -46,40 +46,14 @@ std::unordered_map<chip_id_t, std::vector<CoreCoord>> get_ethernet_cores_grouped
     return tt::tt_metal::MetalContext::instance().get_cluster().get_ethernet_cores_grouped_by_connected_chips(chip_id);
 }
 
-// Get the physical chip ids for a mesh
-std::uint32_t get_ubb_asic_id(chip_id_t physical_chip_id) {
-    auto unique_chip_id =
-        tt::tt_metal::MetalContext::instance().get_cluster().get_unique_chip_ids().at(physical_chip_id);
-    return ((unique_chip_id >> 56) & 0xFF);
-}
-
-bool is_external_ubb_cable(chip_id_t physical_chip_id, CoreCoord eth_core) {
-    auto chan_id = tt::tt_metal::MetalContext::instance()
-                       .get_cluster()
-                       .get_soc_desc(physical_chip_id)
-                       .logical_eth_core_to_chan_map.at(eth_core);
-    auto ubb_asic_id = get_ubb_asic_id(physical_chip_id);
-    bool is_external_cable = false;
-    if (ubb_asic_id == 1) {
-        // UBB 1 has external cables on channesl 0-7
-        is_external_cable = (chan_id >= 0 and chan_id <= 7);
-    } else if (ubb_asic_id >= 2 and ubb_asic_id <= 4) {
-        // UBB 2 to 4 has external cables on channesl 0-3
-        is_external_cable = (chan_id >= 0 and chan_id <= 3);
-    } else if (ubb_asic_id == 5) {
-        // UBB 5 has external cables on channesl 4-7
-        is_external_cable = (chan_id >= 4 and chan_id <= 7);
-    }
-    return is_external_cable;
-}
-
 bool is_chip_on_edge_of_mesh(
     chip_id_t physical_chip_id,
     int num_ports_per_side,
     const std::unordered_map<chip_id_t, std::vector<CoreCoord>>& ethernet_cores_grouped_by_connected_chips) {
+    const auto& cluster = tt::tt_metal::MetalContext::instance().get_cluster();
     // Chip is on edge if it does not have full connections to four sides
-    if (tt::tt_metal::MetalContext::instance().get_cluster().get_board_type(physical_chip_id) == BoardType::UBB) {
-        auto ubb_asic_id = get_ubb_asic_id(physical_chip_id);
+    if (cluster.get_board_type(physical_chip_id) == BoardType::UBB) {
+        auto ubb_asic_id = cluster.get_ubb_asic_id(physical_chip_id);
         return (ubb_asic_id >= 2) and (ubb_asic_id <= 5);
     } else {
         int i = 0;
@@ -96,8 +70,9 @@ bool is_chip_on_corner_of_mesh(
     chip_id_t physical_chip_id,
     int num_ports_per_side,
     const std::unordered_map<chip_id_t, std::vector<CoreCoord>>& ethernet_cores_grouped_by_connected_chips) {
-    if (tt::tt_metal::MetalContext::instance().get_cluster().get_board_type(physical_chip_id) == BoardType::UBB) {
-        auto ubb_asic_id = get_ubb_asic_id(physical_chip_id);
+    const auto& cluster = tt::tt_metal::MetalContext::instance().get_cluster();
+    if (cluster.get_board_type(physical_chip_id) == BoardType::UBB) {
+        auto ubb_asic_id = cluster.get_ubb_asic_id(physical_chip_id);
         return (ubb_asic_id == 1);
     } else {
         // Chip is a corner if it has exactly 2 fully connected sides
@@ -250,10 +225,8 @@ std::vector<chip_id_t> ControlPlane::get_mesh_physical_chip_ids(
         bool is_ubb = cluster.get_board_type(current_chip_id) == BoardType::UBB;
         for (const auto& [connected_chip_id, eth_ports] : eth_links) {
             // Do not include any corner to corner links on UBB
-            if (is_ubb) {
-                if (is_external_ubb_cable(current_chip_id, eth_ports[0])) {
-                    continue;
-                }
+            if (is_ubb && cluster.is_external_cable(current_chip_id, eth_ports[0])) {
+                continue;
             }
             if (eth_ports.size() >= num_ports_per_side) {
                 if (visited_physical_chips.find(connected_chip_id) == visited_physical_chips.end()) {
@@ -321,7 +294,7 @@ std::vector<chip_id_t> ControlPlane::get_mesh_physical_chip_ids(
             bool found_chip = false;
             bool is_ubb = cluster.get_board_type(physical_chip_id_from_north) == BoardType::UBB;
             for (const auto& [connected_chip_id, eth_ports] : eth_links_grouped_by_connected_chips) {
-                if (is_ubb and is_external_ubb_cable(physical_chip_id_from_north, eth_ports[0])) {
+                if (is_ubb && cluster.is_external_cable(physical_chip_id_from_north, eth_ports[0])) {
                     continue;
                 }
                 if (visited_physical_chips.find(connected_chip_id) == visited_physical_chips.end() and

tt_metal/llrt/tt_cluster.cpp

@@ -1422,6 +1422,39 @@ void Cluster::initialize_control_plane() {
     global_control_plane_ = std::make_unique<tt::tt_fabric::GlobalControlPlane>(mesh_graph_desc_path.string());
 }
 
+std::uint32_t Cluster::get_ubb_asic_id(chip_id_t physical_chip_id) const {
+    auto unique_chip_id = this->get_unique_chip_ids().at(physical_chip_id);
+    return ((unique_chip_id >> 56) & 0xFF);
+}
+
+bool Cluster::is_external_cable(chip_id_t physical_chip_id, CoreCoord eth_core) const {
+    auto chan_id = this->get_soc_desc(physical_chip_id).logical_eth_core_to_chan_map.at(eth_core);
+    bool is_external_cable = false;
+    auto board_type = this->get_board_type(physical_chip_id);
+    if (board_type == BoardType::UBB) {
+        auto ubb_asic_id = get_ubb_asic_id(physical_chip_id);
+        if (ubb_asic_id == 1) {
+            // UBB 1 has external cables on channels 0-7
+            is_external_cable = (chan_id >= 0 and chan_id <= 7);
+        } else if (ubb_asic_id >= 2 and ubb_asic_id <= 4) {
+            // UBB 2 to 4 has external cables on channels 0-3
+            is_external_cable = (chan_id >= 0 and chan_id <= 3);
+        } else if (ubb_asic_id == 5) {
+            // UBB 5 has external cables on channels 4-7
+            is_external_cable = (chan_id >= 4 and chan_id <= 7);
+        }
+    } else if (board_type == BoardType::N300) {
+        // N300 has external cables on channels 8-9 on MMIO chips and channels 0-1 on non-MMIO chips
+        auto mmio_device_id = this->get_associated_mmio_device(physical_chip_id);
+        if (mmio_device_id == physical_chip_id) {
+            is_external_cable = (chan_id != 8 and chan_id != 9);
+        } else {
+            is_external_cable = (chan_id != 0 and chan_id != 1);
+        }
+    }
+    return is_external_cable;
+}
+
 }  // namespace tt
 
 std::ostream &operator<<(std::ostream &os, tt_target_dram const &dram) {

tt_metal/llrt/tt_cluster.hpp

@@ -340,6 +340,12 @@ class Cluster {
 
     const std::unordered_map<CoreCoord, int32_t>& get_virtual_routing_to_profiler_flat_id(chip_id_t chip_id) const;
 
+    std::uint32_t get_ubb_asic_id(chip_id_t physical_chip_id) const;
+
+    // TODO: move to separate system descriptor class
+    // return enum for connection type, Internal, QSFP, Other, Unknown
+    bool is_external_cable(chip_id_t physical_chip_id, CoreCoord eth_core) const;
+
 private:
     void detect_arch_and_target();
     void generate_cluster_descriptor();