Skip to content

Blackhole p100/p150 support #2

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Open
wants to merge 1 commit into
base: metalium-support
Choose a base branch
from
Open

Blackhole p100/p150 support #2

Changes from all commits
Commits
Show all changes
1 commit
Select commit
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
9 changes: 6 additions & 3 deletions ggml/src/ggml-metalium/ggml-metalium.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -224,7 +224,7 @@ static size_t g_metalium_base_offset = 0;

static tt::tt_metal::DataType ggml2tt_type_internal(ggml_type ggtype, tt::ARCH arch) {
// This table is consulted to map GGML types to TT types dueing tensor creation
if(arch == tt::ARCH::WORMHOLE_B0) {
if(arch == tt::ARCH::WORMHOLE_B0 || arch == tt::ARCH::BLACKHOLE) {
static constexpr std::array<tt::tt_metal::DataType, GGML_TYPE_COUNT> table = {
/*GGML_TYPE_F32 = */ tt::tt_metal::DataType::BFLOAT16,
/*GGML_TYPE_F16 = */ tt::tt_metal::DataType::BFLOAT16,
Expand Down Expand Up @@ -2573,7 +2573,10 @@ static std::string identify_tensotrrent_device(const ttnn::IDevice* device)
{
auto grid_size = device->compute_with_storage_grid_size();
// TODO: Support mesh configurations
if(device->arch() == tt::ARCH::WORMHOLE_B0) {
if (device->arch() == tt::ARCH::BLACKHOLE) {
return grid_size.x * grid_size.y == 140 ? "Tenstorrent Blackhole p150" : "Tenstorrent Blackhole p100";
}
if (device->arch() == tt::ARCH::WORMHOLE_B0) {
if(grid_size.x == 8 && grid_size.y == 7) {
return "Tenstorrent Wormhole n300";
}
Expand Down Expand Up @@ -2613,7 +2616,7 @@ GGML_BACKEND_API ggml_backend_reg_t ggml_backend_metalium_reg()
ttnn::enable_program_cache(*device);
}
// Limit device support to the ones I own (GS is removed as TTNN dropped support)
GGML_ASSERT(device->arch() == tt::ARCH::WORMHOLE_B0);
GGML_ASSERT(device->arch() == tt::ARCH::WORMHOLE_B0 || device->arch() == tt::ARCH::BLACKHOLE);

dev_ctx->device = device;
dev_ctx->device_id = device_id;
Expand Down