Skip to content

Commit

Permalink
Temporary Workaround for tt-metal Segfaults during teardown (#408)
Browse files Browse the repository at this point in the history
 - A hack, in createBufferFromTensorRef(), remove when proper
   bug fix is made in tt-metal and propagates here.

 - Cannot remove UNSUPPORTED on TTMetal/to_layout.mlir yet because
   CI mixes ttnn/ttmetal binaries fine, but closing device fails,
   need some kind of ttrt fix
  • Loading branch information
kmabeeTT committed Aug 29, 2024
1 parent 25d8e10 commit e2aff3c
Showing 1 changed file with 13 additions and 0 deletions.
13 changes: 13 additions & 0 deletions runtime/include/tt/runtime/detail/ttmetal.h
Original file line number Diff line number Diff line change
Expand Up @@ -159,6 +159,19 @@ createBufferFromTensorRef(::tt::tt_metal::Device *device,
std::shared_ptr<::tt::tt_metal::Buffer> buffer =
::tt::tt_metal::CreateBuffer(shardedBufferConfig);
assert(tensorRef->address());

// Issue #408: Temporary Hack, remove when fix available.
// Update tt-metal BUFFER_MAP with updated address and remove
// entry for original alloc'd address.
auto &buffer_map = tt::tt_metal::detail::BUFFER_MAP;
auto map_copy = buffer_map.value();
auto old_key = std::make_tuple(device->id(), buffer->address());
if (auto it = map_copy.find(old_key); it != map_copy.end()) {
auto new_key = std::make_tuple(device->id(), tensorRef->address());
buffer_map.insert(new_key, it->second);
buffer_map.erase(old_key);
}

buffer->set_address(tensorRef->address());
return buffer;
}
Expand Down

0 comments on commit e2aff3c

Please sign in to comment.