-
Notifications
You must be signed in to change notification settings - Fork 104
Commit
This commit does not belong to any branch on this repository, and may belong to a fork outside of the repository.
LightMetal - Initial Host API + Device Capture infra/library and unit…
… tests (#17039) - This is round 5, builds upon previous 4 merges for LightMetal that brought flatbuffer cmake/infra, begin/end APIs, LoadTrace() API, flatbuffer/schema serialization/deserialization - This adds light-metal Capture support for and instruments with LIGHT_METAL_TRACE_FUNCTION_CALL() and LIGHT_METAL_TRACE_FUNCTION_ENTRY() to many popular (not exuahstive) APIs used by unit tests. The former TRACE_FUNCTION_ENTRY() is more recent, used to protect against host APIs recursively calling other host APIs (only trace top most level). Two macros not always called back-to-back. - Support Capture/Replay of the following ~14 host APIs EnqueueTrace(), ReplayTrace(), ReleaseTrace() CreateBuffer(), EnqueueWriteBuffer(), EnqueueReadBuffer(), DeallocateBuffer CreateKernel(), CreateCircularBuffer() SetRuntimeArgs(uint32) SetRuntimeArgs(Kernel,RuntimeArgs) CreateProgram(), EnqueueProgram() Finish() - During capture, complex objects like Programs, Kernels, Buffers, CBHandle are assigned unique global_id, and referred to by their global_id in capture when used by functions - When "Metal Trace" is enabled, don't capture EnqueueProgram(), instead inject ReplayTrace(), would be used alongside LoadTrace() - Can be optionally disabled at compile time using build_metal.sh --disable-light-metal-trace which will set C++ define TT_ENABLE_LIGHT_METAL_TRACE=0 (trace functions become NOP) - New Verif APIs LightMetalCompareToCapture() / LightMetalCompareToGolden(). Put them in lightmetal_capture_utils.hpp instead of host_api.hpp since they are purely used at capture time, and not worthy enough to be inside host_api.h since just for verif - Test fixture runs capture-only right now, will automatically run binary once replay support is merged next.
- Loading branch information
Showing
21 changed files
with
1,522 additions
and
18 deletions.
There are no files selected for viewing
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
| Original file line number | Diff line number | Diff line change |
|---|---|---|
| @@ -0,0 +1,23 @@ | ||
| set(UNIT_TESTS_LIGHTMETAL_SRC ${CMAKE_CURRENT_SOURCE_DIR}/test_lightmetal.cpp) | ||
|
|
||
| add_executable(unit_tests_lightmetal ${UNIT_TESTS_LIGHTMETAL_SRC}) | ||
| TT_ENABLE_UNITY_BUILD(unit_tests_lightmetal) | ||
|
|
||
| target_link_libraries(unit_tests_lightmetal PUBLIC test_metal_common_libs) | ||
|
|
||
| target_include_directories( | ||
| unit_tests_lightmetal | ||
| PRIVATE | ||
| "$<TARGET_PROPERTY:Metalium::Metal,INCLUDE_DIRECTORIES>" | ||
| ${PROJECT_SOURCE_DIR}/tests | ||
| ${PROJECT_SOURCE_DIR}/tests/tt_metal/tt_metal/common | ||
| ${CMAKE_CURRENT_SOURCE_DIR} | ||
| ${CMAKE_CURRENT_SOURCE_DIR}/common | ||
| ) | ||
|
|
||
| set_target_properties( | ||
| unit_tests_lightmetal | ||
| PROPERTIES | ||
| RUNTIME_OUTPUT_DIRECTORY | ||
| ${PROJECT_BINARY_DIR}/test/tt_metal | ||
| ) |
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
| Original file line number | Diff line number | Diff line change |
|---|---|---|
| @@ -0,0 +1,60 @@ | ||
| // SPDX-FileCopyrightText: © 2025 Tenstorrent Inc. | ||
| // | ||
| // SPDX-License-Identifier: Apache-2.0 | ||
|
|
||
| #pragma once | ||
|
|
||
| #include "dispatch_fixture.hpp" | ||
| #include <tt-metalium/device_impl.hpp> | ||
| #include <tt-metalium/hal.hpp> | ||
| #include <tt-metalium/host_api.hpp> | ||
| #include <tt-metalium/tt_metal.hpp> | ||
| #include <circular_buffer_constants.h> | ||
| #include <tt-metalium/kernel.hpp> | ||
| #include <tt-metalium/tt_backend_api_types.hpp> | ||
| #include "command_queue_fixture.hpp" | ||
| #include <lightmetal_binary.hpp> | ||
|
|
||
| class SingleDeviceLightMetalFixture : public CommandQueueFixture { | ||
| protected: | ||
| std::string trace_bin_path_; | ||
| bool write_bin_to_disk_; | ||
|
|
||
| void SetUp() override { | ||
| this->validate_dispatch_mode(); | ||
| this->arch_ = tt::get_arch_from_string(tt::test_utils::get_umd_arch_name()); | ||
| } | ||
|
|
||
| void CreateDevice( | ||
| const size_t trace_region_size, const bool replay_binary = true, const std::string trace_bin_path = "") { | ||
| // Skip writing to disk by default, unless user sets env var for local testing | ||
| write_bin_to_disk_ = tt::parse_env("LIGHTMETAL_SAVE_BINARY", false); | ||
|
|
||
| // If user didn't provide a specific trace bin path, set a default here based on test name | ||
| if (trace_bin_path == "") { | ||
| const auto test_info = ::testing::UnitTest::GetInstance()->current_test_info(); | ||
| auto trace_filename = test_info ? std::string(test_info->name()) + ".bin" : "lightmetal_trace.bin"; | ||
| this->trace_bin_path_ = "/tmp/" + trace_filename; | ||
| } | ||
|
|
||
| this->create_device(trace_region_size); | ||
| LightMetalBeginCapture(); | ||
| } | ||
|
|
||
| // End light metal tracing, write to optional filename and optionally run from binary blob | ||
| void TearDown() override { | ||
| LightMetalBinary binary = LightMetalEndCapture(); | ||
|
|
||
| if (binary.is_empty()) { | ||
| FAIL() << "Light Metal Binary is empty for test, unexpected."; | ||
| } | ||
| if (write_bin_to_disk_ && !this->trace_bin_path_.empty() && !binary.is_empty()) { | ||
| log_info(tt::LogTest, "Writing light metal binary {} bytes to {}", binary.size(), this->trace_bin_path_); | ||
| binary.save_to_file(this->trace_bin_path_); | ||
| } | ||
|
|
||
| if (!this->IsSlowDispatch()) { | ||
| tt::tt_metal::CloseDevice(this->device_); | ||
| } | ||
| } | ||
| }; |
Oops, something went wrong.