Add support for new logical sharding + alignment in TensorLayout and tensor creation #14771
Conversation
![github-actions[bot]](https://avatars.githubusercontent.com/in/15368?s=60&v=4){kind=small}
This comment was marked as outdated.
This comment was marked as outdated.
Sorry, something went wrong.
TT-BrianLiu
force-pushed
the
bliu/main2
branch
from
e7375dd
to
4d32ca5
Compare
TT-BrianLiu
requested review from
ayerofieiev-tt,
dmakoviichuk-tt,
rfurko-tt,
cfjchu,
razorback3 and
dongjin-na
as code owners
TT-BrianLiu
force-pushed
the
bliu/main2
branch
2 times, most recently
from
dc158eb
to
8d7f1b4
Compare
This comment was marked as outdated.
This comment was marked as outdated.
Sorry, something went wrong.
tests/ttnn/unit_tests/gtests/tensor/test_sharding_with_alignment.cpp
Outdated
Show resolved
Hide resolved
This comment was marked as outdated.
This comment was marked as outdated.
Sorry, something went wrong.
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Alignment" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/layout/alignment.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::PageConfig" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/layout/page_config.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Layout" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/enum_types.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Layout::ROW_MAJOR" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/enum_types.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::TensorMemoryLayout" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::TensorMemoryLayout::WIDTH_SHARDED" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::BufferType" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::BufferType::L1" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::ShardSpec" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| .memory_layout = TensorMemoryLayout::WIDTH_SHARDED, | ||
| .buffer_type = BufferType::L1, | ||
| .shard_spec = ShardSpec{ | ||
| num_cores_to_corerange_set(tt::div_up(5, 1), grid_size, /*row_wise=*/true), |
There was a problem hiding this comment.
use of undeclared identifier num_cores_to_corerange_set; did you mean num_cores_to_corerangeset?
| num_cores_to_corerange_set(tt::div_up(5, 1), grid_size, /*row_wise=*/true), | |
| num_cores_to_corerangeset(tt::div_up(5, 1), grid_size, /*row_wise=*/true), |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::div_up" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "common/math.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::ShardOrientation" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::ShardOrientation::ROW_MAJOR" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Size" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/layout/size.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Size" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/layout/size.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "ttnn/tensor/layout/tensor_layout.hpp" | ||
| #include "ttnn/tensor/tensor.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::SimpleShape" is directly included
| #include "ttnn/tensor/tensor.hpp" | |
| #include "ttnn/tensor/shape/shape.hpp" | |
| #include "ttnn/tensor/tensor.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "std::nullopt" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include <optional> | |
| #include "common_tensor_test_utils.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Alignment" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/layout/alignment.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::PageConfig" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/layout/page_config.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Layout" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/enum_types.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Layout::ROW_MAJOR" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/enum_types.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::TensorMemoryLayout" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::TensorMemoryLayout::WIDTH_SHARDED" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::BufferType" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::BufferType::L1" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::ShardSpec" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| .memory_layout = TensorMemoryLayout::WIDTH_SHARDED, | ||
| .buffer_type = BufferType::L1, | ||
| .shard_spec = ShardSpec{ | ||
| num_cores_to_corerange_set(tt::div_up(5, 4), grid_size, /*row_wise=*/true), |
There was a problem hiding this comment.
use of undeclared identifier num_cores_to_corerange_set; did you mean num_cores_to_corerangeset?
| num_cores_to_corerange_set(tt::div_up(5, 4), grid_size, /*row_wise=*/true), | |
| num_cores_to_corerangeset(tt::div_up(5, 4), grid_size, /*row_wise=*/true), |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::div_up" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "common/math.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::ShardOrientation" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::ShardOrientation::ROW_MAJOR" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Size" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/layout/size.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Size" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/layout/size.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "ttnn/tensor/layout/tensor_layout.hpp" | ||
| #include "ttnn/tensor/tensor.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::SimpleShape" is directly included
| #include "ttnn/tensor/tensor.hpp" | |
| #include "ttnn/tensor/shape/shape.hpp" | |
| #include "ttnn/tensor/tensor.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "std::nullopt" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include <optional> | |
| #include "common_tensor_test_utils.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Alignment" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/layout/alignment.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::PageConfig" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/layout/page_config.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Layout" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/enum_types.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Layout::ROW_MAJOR" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/enum_types.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::TensorMemoryLayout" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::TensorMemoryLayout::INTERLEAVED" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::BufferType" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::BufferType::L1" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "std::nullopt" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include <optional> | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "std::nullopt" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include <optional> | |
| #include "common_tensor_test_utils.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Size" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/layout/size.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "ttnn/tensor/layout/tensor_layout.hpp" | ||
| #include "ttnn/tensor/tensor.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::SimpleShape" is directly included
| #include "ttnn/tensor/tensor.hpp" | |
| #include "ttnn/tensor/shape/shape.hpp" | |
| #include "ttnn/tensor/tensor.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Size" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/layout/size.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Alignment" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/layout/alignment.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::PageConfig" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/layout/page_config.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Layout" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/enum_types.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Layout::TILE" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/enum_types.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "common_tensor_test_utils.hpp" | ||
| #include "gtest/gtest.h" | ||
| #include "host_api.hpp" | ||
| #include "tt_metal/common/logger.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Tile" is directly included
| #include "tt_metal/common/logger.hpp" | |
| #include "tile/tile.hpp" | |
| #include "tt_metal/common/logger.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::TensorMemoryLayout" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::TensorMemoryLayout::BLOCK_SHARDED" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::BufferType" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::BufferType::L1" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::ShardSpec" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| .memory_layout = TensorMemoryLayout::BLOCK_SHARDED, | ||
| .buffer_type = BufferType::L1, | ||
| .shard_spec = ShardSpec{ | ||
| num_cores_to_corerange_set(tt::div_up(8 * 36, 48) * tt::div_up(32, 10), grid_size, /*row_wise=*/true), |
There was a problem hiding this comment.
use of undeclared identifier num_cores_to_corerange_set; did you mean num_cores_to_corerangeset?
| num_cores_to_corerange_set(tt::div_up(8 * 36, 48) * tt::div_up(32, 10), grid_size, /*row_wise=*/true), | |
| num_cores_to_corerangeset(tt::div_up(8 * 36, 48) * tt::div_up(32, 10), grid_size, /*row_wise=*/true), |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::div_up" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "common/math.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::div_up" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "common/math.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::ShardOrientation" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::ShardOrientation::ROW_MAJOR" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Size" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/layout/size.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Size" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/layout/size.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "ttnn/tensor/layout/tensor_layout.hpp" | ||
| #include "ttnn/tensor/tensor.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::SimpleShape" is directly included
| #include "ttnn/tensor/tensor.hpp" | |
| #include "ttnn/tensor/shape/shape.hpp" | |
| #include "ttnn/tensor/tensor.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Size" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/layout/size.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Alignment" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/layout/alignment.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::PageConfig" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/layout/page_config.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Layout" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/enum_types.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Layout::ROW_MAJOR" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/enum_types.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::TensorMemoryLayout" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::TensorMemoryLayout::BLOCK_SHARDED" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::BufferType" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::BufferType::L1" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::ShardSpec" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| .memory_layout = TensorMemoryLayout::BLOCK_SHARDED, | ||
| .buffer_type = BufferType::L1, | ||
| .shard_spec = ShardSpec{ | ||
| num_cores_to_corerange_set(tt::div_up(2 * 10, 5) * tt::div_up(5, 2), grid_size, /*row_wise=*/true), |
There was a problem hiding this comment.
use of undeclared identifier num_cores_to_corerange_set; did you mean num_cores_to_corerangeset?
| num_cores_to_corerange_set(tt::div_up(2 * 10, 5) * tt::div_up(5, 2), grid_size, /*row_wise=*/true), | |
| num_cores_to_corerangeset(tt::div_up(2 * 10, 5) * tt::div_up(5, 2), grid_size, /*row_wise=*/true), |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::div_up" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "common/math.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::div_up" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "common/math.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::ShardOrientation" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| @@ -1,34 +1,18 @@ | |||
| // SPDX-FileCopyrightText: © 2024 Tenstorrent Inc. | |||
| // SPDX-License-Identifier: Apache-2.0 | |||
|
|
|||
| #include <host_api.hpp> | |||
| #include <ttnn/tensor/tensor.hpp> | |||
|
|
|||
| #include "common_tensor_test_utils.hpp" | |||
There was a problem hiding this comment.
no header providing "tt::tt_metal::ShardOrientation::ROW_MAJOR" is directly included
| #include "common_tensor_test_utils.hpp" | |
| #include "buffers/buffer_constants.hpp" | |
| #include "common_tensor_test_utils.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Size" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/layout/size.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
| #include "tt_metal/common/logger.hpp" | ||
| #include "tt_metal/common/work_split.hpp" | ||
| #include "ttnn/async_runtime.hpp" | ||
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
There was a problem hiding this comment.
no header providing "tt::tt_metal::Size" is directly included
| #include "ttnn/tensor/layout/tensor_layout.hpp" | |
| #include "ttnn/tensor/layout/size.hpp" | |
| #include "ttnn/tensor/layout/tensor_layout.hpp" |
TT-BrianLiu
force-pushed
the
bliu/main2
branch
from
8d7f1b4
to
1f82c5b
Compare
TT-BrianLiu
requested review from
tt-dma,
tt-asaigal and
ubcheema
as code owners
TT-BrianLiu
force-pushed
the
bliu/main2
branch
7 times, most recently
from
b9af550
to
7837fde
Compare
…out and tensor creation
- Add ShardMode enum to specify shard shape in shard spec as either physical or logical
* ShardMode::PHYSICAL: This is current behaviour that we will deprecate!
** It is less expressive than using shard shape as logical (ie. must be tile aligned for TILE layout etc...)
** It fundamentally operates on padded shape and is confusing and incompatible with logical shape
* ShardMode::LOGICAL: Shard shape cuts 2D logical shape and each shard is aligned after
** Without alignment restrictions, you can cut 2D logical shape more arbitrarily
** Existing sharding can be switched over to this entirely (just need codeowners to help out and flip...)
* Default everywhere will be ShardMode::PHYSICAL with a warning message
- Switch tests/ttnn/unit_tests/operations/test_paged_update_cache.py to use logical shard shape as an example
* Introduce tensor.logical_volume() (as opposed to tensor.volume() which returns physical volume based on padded shape)
* TODO: Rename volume() -> physical_volume() and logical_volume() -> volume()
- Add new c++ tests to test tensor creation with logical shard shape + alignment
* IMPORTANT: Need to update host data manipulation to be aware of new logical sharding for use from python!
To support these changes, some changes to TensorLayout:
- Make private TensorLayout constructor with alignment public with these changes:
* legacyShapeToAlignment will try to return 2D alignment if possible (ie. only padding on height/width)
** Goal is to transition alignment to be 2D only if we remove poor use cases of padding on non-height/width dims
* legacyShapeToAlignment is only expected to be used for ShardMode::PHYSICAL and uses default alignment for sharded tensors
** Before interleaved or sharded will just use padded shape for alignment
** One exception is for row major sharded tensors where we use shard width if shape is padded;
otherwise, we only take shard width for BLOCK/WIDTH sharded cases and original physical shape for HEIGHT sharded
* legacyShapeToAlignment (and alignment in general) will work iff there is only padding on height and/or width
** IMPORTANT: This means we are expecting tensors with arbitrary padding along non-height/width to be interleaved only!
- If ShardMode::LOGICAL:
* In TensorLayout::compute_shard_spec_buffer, calculate physical shard shape based on shard shape + alignment
* In TensorLayout::compute_physical_shape, calculate physical shape based on number of logical shards
- Clean up handling of sharded tensors and error messages in ttnn/cpp/ttnn/tensor/layout/page_config.cpp
- Add Size constructor for std::array<uint32_t, 2>
TT-BrianLiu
changed the title
TT-BrianLiu
force-pushed
the
bliu/main2
branch
from
7837fde
to
f7354bf
Compare
| @@ -24,6 +24,8 @@ class TensorLayout { | |||
| public: | |||
| TensorLayout(DataType dtype, const PageConfig& page_config, const MemoryConfig& memory_config); | |||
|
|
|||
| TensorLayout(DataType dtype, const PageConfig& page_config, const MemoryConfig& memory_config, const Alignment& alignment); | |||
There was a problem hiding this comment.
Let's discuss this. I explicitly hid this constructor because alignment is a detail of implementation which we did not want to expose. 2D layout can be automatically deduced. ND is confusing and was only necessary to encode bad padding.
| } | ||
| return Alignment{}; | ||
| } else { | ||
| if (alignment_can_be_2D) { |
There was a problem hiding this comment.
@TT-BrianLiu Do we need this check? The code inside seems very similar, maybe we can unify it more?
There was a problem hiding this comment.
I want to predetermine if alignment can be 2D and keep values size 2 or less if possible
| if (memory_config_.shard_spec.has_value() and memory_config_.shard_spec.value().mode == ShardMode::LOGICAL) { | ||
| // Iterate dims in reverse order | ||
| for (int i = -1; i >= -rank; --i) { | ||
| auto& dim = i == -1 ? width : height; | ||
| dim *= shape[i]; | ||
| } |
There was a problem hiding this comment.
Can we please change this loop to be more straightforward?
It should be equivalent to:
size_t width = shape[-1];
size_t height = 1;
for (size_t i = 0; i + 1 < rank; i++) {
dim *= shape[i];
}
There was a problem hiding this comment.
For sharded, it works without much code change. For interleaved, it goes in reverse because alignment is matched that way.
| // Iterate dims in reverse order and ensure alignment | ||
| // Even tensor of rank 0 or 1 must be aligned (to Tile / Page / Shard) | ||
| for (int i = -1; i >= -max_rank; --i) { | ||
| auto& dim = i == -1 ? width : height; |
There was a problem hiding this comment.
Can we make it more straightforward if we handle case of i = -1 outside of the loop?
There was a problem hiding this comment.
code can be shared with alignment logic. Do we want to duplicate?
Ticket
Link to Github Issue
Problem description
To move away from padded shape, we are transitioning sharding to be logical. This is more expressive than sharding with physical shapes that need to be TILE aligned for example.
What's changed
To support these changes, some changes to TensorLayout:
otherwise, we only take shard width for BLOCK/WIDTH sharded cases and original physical shape for HEIGHT sharded
Checklist