From ee9faa4a59f8a21f1477b863d6fc345184ef94fa Mon Sep 17 00:00:00 2001 From: Austin Ho Date: Fri, 6 Dec 2024 21:10:58 +0000 Subject: [PATCH] #0: Swap unordered_map for map in global circular buffer since this object needs to be hashed --- .../tt_metal/api/test_global_circular_buffers.cpp | 10 +++++----- .../test_dram_read_remote_cb.cpp | 4 ++-- .../test_remote_cb_sync_matmul.cpp | 4 ++-- tt_metal/impl/buffers/global_circular_buffer.cpp | 4 ++-- tt_metal/impl/buffers/global_circular_buffer.hpp | 6 +++--- tt_metal/include/tt_metal/global_circular_buffer.hpp | 2 +- tt_metal/tt_metal.cpp | 2 +- ttnn/cpp/pybind11/global_circular_buffer.cpp | 4 ++-- ttnn/cpp/ttnn/global_circular_buffer.cpp | 4 ++-- ttnn/cpp/ttnn/global_circular_buffer.hpp | 4 ++-- 10 files changed, 22 insertions(+), 22 deletions(-) diff --git a/tests/tt_metal/tt_metal/api/test_global_circular_buffers.cpp b/tests/tt_metal/tt_metal/api/test_global_circular_buffers.cpp index f07f505ff0e1..6b6835e078ac 100644 --- a/tests/tt_metal/tt_metal/api/test_global_circular_buffers.cpp +++ b/tests/tt_metal/tt_metal/api/test_global_circular_buffers.cpp @@ -21,7 +21,7 @@ TEST_F(DispatchFixture, TensixCreateGlobalCircularBuffers) { auto device = devices_[0]; { - std::unordered_map sender_receiver_core_mapping; + std::map sender_receiver_core_mapping; sender_receiver_core_mapping[CoreCoord(0, 0)] = cores; auto global_cb = tt::tt_metal::v1::experimental::CreateGlobalCircularBuffer( device, sender_receiver_core_mapping, 3200, tt::tt_metal::BufferType::L1); @@ -29,7 +29,7 @@ TEST_F(DispatchFixture, TensixCreateGlobalCircularBuffers) { auto config_address = global_cb->config_address(); } { - std::unordered_map sender_receiver_core_mapping; + std::map sender_receiver_core_mapping; sender_receiver_core_mapping[CoreCoord(0, 0)] = cores; sender_receiver_core_mapping[CoreCoord(1, 1)] = cores3; // sender receiver cores overlap @@ -39,7 +39,7 @@ TEST_F(DispatchFixture, TensixCreateGlobalCircularBuffers) { std::exception); } { - std::unordered_map sender_receiver_core_mapping; + std::map sender_receiver_core_mapping; sender_receiver_core_mapping[CoreCoord(0, 0)] = cores; sender_receiver_core_mapping[CoreCoord(0, 1)] = cores2; // receiver cores overlap @@ -60,11 +60,11 @@ TEST_F(DispatchFixture, TensixProgramGlobalCircularBuffers) { tt::DataFormat tile_format = tt::DataFormat::Float16_b; auto all_cores = sender_cores.merge(receiver_cores).merge(dummy_receiver_cores); auto device = devices_[0]; - std::unordered_map sender_receiver_core_mapping; + std::map sender_receiver_core_mapping; sender_receiver_core_mapping[CoreCoord(0, 0)] = receiver_cores; auto global_cb = tt::tt_metal::v1::experimental::CreateGlobalCircularBuffer( device, sender_receiver_core_mapping, 3200, tt::tt_metal::BufferType::L1); - std::unordered_map dummy_sender_receiver_core_mapping; + std::map dummy_sender_receiver_core_mapping; dummy_sender_receiver_core_mapping[CoreCoord(0, 0)] = dummy_receiver_cores; auto dummy_global_cb = tt::tt_metal::v1::experimental::CreateGlobalCircularBuffer( device, dummy_sender_receiver_core_mapping, 3200, tt::tt_metal::BufferType::L1); diff --git a/tests/tt_metal/tt_metal/perf_microbenchmark/10_dram_read_remote_cb_sync/test_dram_read_remote_cb.cpp b/tests/tt_metal/tt_metal/perf_microbenchmark/10_dram_read_remote_cb_sync/test_dram_read_remote_cb.cpp index 5480b3144c0d..7d5f6a14f3c5 100644 --- a/tests/tt_metal/tt_metal/perf_microbenchmark/10_dram_read_remote_cb_sync/test_dram_read_remote_cb.cpp +++ b/tests/tt_metal/tt_metal/perf_microbenchmark/10_dram_read_remote_cb_sync/test_dram_read_remote_cb.cpp @@ -85,7 +85,7 @@ create_programs( tt_metal::Device* device, const CoreRangeSet& dram_reader_core, const CoreRangeSet& l1_receiver_cores, - const std::unordered_map& sender_receiver_core_mapping, + const std::map& sender_receiver_core_mapping, const uint32_t& single_tile_size, const tt::DataFormat& tile_format, uint32_t k, @@ -721,7 +721,7 @@ int main(int argc, char** argv) { l1_receiver_core_coord_range = CoreRange{CoreCoord{1, 0}, CoreCoord{num_receivers, 0}}; } CoreRangeSet l1_receiver_core{std::set{l1_receiver_core_coord_range}}; - std::unordered_map sender_receiver_core_mapping; + std::map sender_receiver_core_mapping; sender_receiver_core_mapping[dram_reader_core_coord] = l1_receiver_core; if (use_sub_devices) { SubDevice sender_sub_device = SubDevice(std::array{dram_reader_core}); diff --git a/tests/tt_metal/tt_metal/perf_microbenchmark/11_remote_cb_sync_matmul_single_core/test_remote_cb_sync_matmul.cpp b/tests/tt_metal/tt_metal/perf_microbenchmark/11_remote_cb_sync_matmul_single_core/test_remote_cb_sync_matmul.cpp index 1d467d9d47d4..8680264b1022 100644 --- a/tests/tt_metal/tt_metal/perf_microbenchmark/11_remote_cb_sync_matmul_single_core/test_remote_cb_sync_matmul.cpp +++ b/tests/tt_metal/tt_metal/perf_microbenchmark/11_remote_cb_sync_matmul_single_core/test_remote_cb_sync_matmul.cpp @@ -100,7 +100,7 @@ create_programs( tt_metal::Device* device, const CoreRangeSet& dram_reader_core, const CoreRangeSet& l1_receiver_cores, - const std::unordered_map& sender_receiver_core_mapping, + const std::map& sender_receiver_core_mapping, const uint32_t& single_tile_size, const tt::DataFormat& tile_format, uint32_t m, @@ -700,7 +700,7 @@ int main(int argc, char** argv) { l1_receiver_core_coord_range = CoreRange{CoreCoord{1, 0}, CoreCoord{num_receivers, 0}}; } CoreRangeSet l1_receiver_core{std::set{l1_receiver_core_coord_range}}; - std::unordered_map sender_receiver_core_mapping; + std::map sender_receiver_core_mapping; sender_receiver_core_mapping[dram_reader_core_coord] = l1_receiver_core; if (use_sub_devices) { SubDevice sender_sub_device = SubDevice(std::array{dram_reader_core}); diff --git a/tt_metal/impl/buffers/global_circular_buffer.cpp b/tt_metal/impl/buffers/global_circular_buffer.cpp index 4d438e91fcce..9de9819167da 100644 --- a/tt_metal/impl/buffers/global_circular_buffer.cpp +++ b/tt_metal/impl/buffers/global_circular_buffer.cpp @@ -25,7 +25,7 @@ namespace experimental { GlobalCircularBuffer::GlobalCircularBuffer( Device* device, - const std::unordered_map& sender_receiver_core_mapping, + const std::map& sender_receiver_core_mapping, uint32_t size, BufferType buffer_type) : device_(device), sender_receiver_core_mapping_(sender_receiver_core_mapping), size_(size) { @@ -134,7 +134,7 @@ void GlobalCircularBuffer::setup_cb_buffers(BufferType buffer_type, uint32_t max std::shared_ptr GlobalCircularBuffer::create( Device* device, - const std::unordered_map& sender_receiver_core_mapping, + const std::map& sender_receiver_core_mapping, uint32_t size, BufferType buffer_type) { return std::make_unique(device, sender_receiver_core_mapping, size, buffer_type); diff --git a/tt_metal/impl/buffers/global_circular_buffer.hpp b/tt_metal/impl/buffers/global_circular_buffer.hpp index c263fe47d002..56b1e07c4fad 100644 --- a/tt_metal/impl/buffers/global_circular_buffer.hpp +++ b/tt_metal/impl/buffers/global_circular_buffer.hpp @@ -28,7 +28,7 @@ class GlobalCircularBuffer { public: GlobalCircularBuffer( Device* device, - const std::unordered_map& sender_receiver_core_mapping, + const std::map& sender_receiver_core_mapping, uint32_t size, BufferType buffer_type); @@ -40,7 +40,7 @@ class GlobalCircularBuffer { static std::shared_ptr create( Device* device, - const std::unordered_map& sender_receiver_core_mapping, + const std::map& sender_receiver_core_mapping, uint32_t size, BufferType buffer_type = BufferType::L1); @@ -64,7 +64,7 @@ class GlobalCircularBuffer { std::shared_ptr cb_buffer_; std::shared_ptr cb_config_buffer_; Device* device_; - std::unordered_map sender_receiver_core_mapping_; + std::map sender_receiver_core_mapping_; CoreRangeSet sender_cores_; CoreRangeSet receiver_cores_; CoreRangeSet all_cores_; diff --git a/tt_metal/include/tt_metal/global_circular_buffer.hpp b/tt_metal/include/tt_metal/global_circular_buffer.hpp index 776296a589a4..924c2dd1ab10 100644 --- a/tt_metal/include/tt_metal/global_circular_buffer.hpp +++ b/tt_metal/include/tt_metal/global_circular_buffer.hpp @@ -26,7 +26,7 @@ namespace experimental { */ std::shared_ptr CreateGlobalCircularBuffer( Device* device, - const std::unordered_map& sender_receiver_core_mapping, + const std::map& sender_receiver_core_mapping, uint32_t size, BufferType buffer_type = BufferType::L1); diff --git a/tt_metal/tt_metal.cpp b/tt_metal/tt_metal.cpp index ff1987983e9e..9a2ee9c48756 100644 --- a/tt_metal/tt_metal.cpp +++ b/tt_metal/tt_metal.cpp @@ -1369,7 +1369,7 @@ namespace experimental { std::shared_ptr CreateGlobalCircularBuffer( Device* device, - const std::unordered_map& sender_receiver_core_mapping, + const std::map& sender_receiver_core_mapping, uint32_t size, BufferType buffer_type) { return GlobalCircularBuffer::create(device, sender_receiver_core_mapping, size, buffer_type); diff --git a/ttnn/cpp/pybind11/global_circular_buffer.cpp b/ttnn/cpp/pybind11/global_circular_buffer.cpp index f736ee99781a..3ae2c89e581e 100644 --- a/ttnn/cpp/pybind11/global_circular_buffer.cpp +++ b/ttnn/cpp/pybind11/global_circular_buffer.cpp @@ -19,7 +19,7 @@ void py_module(py::module& module) { // Single Device APIs module.def( "create_global_circular_buffer", - py::overload_cast&, uint32_t, BufferType>( + py::overload_cast&, uint32_t, BufferType>( &create_global_circular_buffer), py::arg("device"), py::arg("sender_receiver_core_mapping"), @@ -38,7 +38,7 @@ void py_module(py::module& module) { // Multi Device APIs module.def( "create_global_circular_buffer", - py::overload_cast&, uint32_t, BufferType>( + py::overload_cast&, uint32_t, BufferType>( &create_global_circular_buffer), py::arg("mesh_device"), py::arg("sender_receiver_core_mapping"), diff --git a/ttnn/cpp/ttnn/global_circular_buffer.cpp b/ttnn/cpp/ttnn/global_circular_buffer.cpp index 7c5967fa3c2c..8b8c9fe452bb 100644 --- a/ttnn/cpp/ttnn/global_circular_buffer.cpp +++ b/ttnn/cpp/ttnn/global_circular_buffer.cpp @@ -19,7 +19,7 @@ MultiDeviceGlobalCircularBuffer::MultiDeviceGlobalCircularBuffer(MeshDevice* mes std::shared_ptr create_global_circular_buffer( Device* device, - const std::unordered_map& sender_receiver_core_mapping, + const std::map& sender_receiver_core_mapping, uint32_t size, BufferType buffer_type) { std::shared_ptr global_cb; @@ -34,7 +34,7 @@ std::shared_ptr create_global_circular_buffer( MultiDeviceGlobalCircularBuffer create_global_circular_buffer( MeshDevice* mesh_device, - const std::unordered_map& sender_receiver_core_mapping, + const std::map& sender_receiver_core_mapping, uint32_t size, BufferType buffer_type) { MultiDeviceGlobalCircularBuffer multi_device_global_cb(mesh_device); diff --git a/ttnn/cpp/ttnn/global_circular_buffer.hpp b/ttnn/cpp/ttnn/global_circular_buffer.hpp index bb84ce3a7ab3..1469f6a1f3c1 100644 --- a/ttnn/cpp/ttnn/global_circular_buffer.hpp +++ b/ttnn/cpp/ttnn/global_circular_buffer.hpp @@ -18,14 +18,14 @@ struct MultiDeviceGlobalCircularBuffer { // Single Device APIs std::shared_ptr create_global_circular_buffer( Device* device, - const std::unordered_map& sender_receiver_core_mapping, + const std::map& sender_receiver_core_mapping, uint32_t size, BufferType buffer_type = BufferType::L1); // Multi Device APIs MultiDeviceGlobalCircularBuffer create_global_circular_buffer( MeshDevice* mesh_device, - const std::unordered_map& sender_receiver_core_mapping, + const std::map& sender_receiver_core_mapping, uint32_t size, BufferType buffer_type = BufferType::L1);