#2934: Make one CommandQueue and one HW CommandQueue (SysmemWriter) per device #4077

abhullar-tt · 2023-11-29T20:00:06Z

Added CommandQueue &detail::GetCommandQueue(Device *device) API which creates a CQ for given device if it doesn't already exist and returns ref to it.

Fast dispatch Enqueue APIs used to accept reference to GLOBAL_CQ but users are expected to pass CQ returned from this API.

Also moved SysmemWriter into tt_metal::Device.

Eventually we can add an API to create command queues to enable multiple SW CQs and for one device all SW CQs will be using the same HW CQ(s)

tt_metal/detail/tt_metal.hpp

tt_metal/llrt/tt_cluster.cpp

abhullar-tt · 2023-12-05T18:29:07Z

Tests added:

DeviceFixture.TestDeviceToHostMemChannelAssignment

checks that each device is assigned to unique host memory channel. There is one hugepage per host mem channel

MultiCommandQueueFixture.TestAccessCommandQueue

checks that we can access a SW CQ for each device that has been initialized

BasicFastDispatchFixture.TestCannotAccessCommandQueueForClosedDevice

opens device -> closes device -> and then ensures we can't access CQ for a closed device

MultiCommandQueueFixture.TestDirectedLoopbackToUniqueHugepage

uses cluster APIs to do directed write and reads from the same address across multiple hugepages to ensure each device hugepage is independent

Additional tests for multi-command queue will be incrementally added when we add additional N300 fast dispatch features

tests/tt_metal/tt_metal/unit_tests/compute/matmul/single_core_matmul_compute.cpp

tests/tt_metal/tt_metal/unit_tests_fast_dispatch/command_queue/test_CommandQueue.cpp

tests/tt_metal/tt_metal/unit_tests_fast_dispatch/common/command_queue_fixture.hpp

tt_metal/impl/dispatch/command_queue_interface.hpp

abhullar-tt · 2023-12-06T17:14:49Z

@tt-rkim this PR can't be merged to main until we enable creating more than 1 hugepage on our machines

tt_metal/detail/tt_metal.hpp

DrJessop · 2023-12-06T23:12:17Z

tt_metal/detail/tt_metal.hpp

        {
-            if (detail::GLOBAL_CQ) {
-                ClearProgramCache(*detail::GLOBAL_CQ);
+            if (std::getenv("TT_METAL_SLOW_DISPATCH_MODE") == nullptr) {


curious whether we should assert otherwise?

I think these APIs get called in slow dispatch mode atm as well

DrJessop · 2023-12-06T23:13:16Z

tt_metal/detail/tt_metal.hpp

+            static std::mutex cq_creation_mutex;
+            {
+                std::lock_guard<std::mutex> lock(cq_creation_mutex);
+                if (not command_queues[id] or (command_queues[id] and command_queues[id]->device != device)) {


why would command_queue[id] ever be initialize and command_queues[id]->device != device?

yeah so originally I only had CQ being created once but then this was causing a bug where it complained that CQ
was using a device that was closed. This case happens when:

In the same process: device = CreateDevice(0); EnqueueX(GetCommandQueue(device), ...) // this will initialize command_queues[device->id] CloseDevice(device) device2 = CreateDevice(0); // technically same device but different object EnqueueX(GetCommandQueue(device2), ...) // command_queues[device2->id] exists because device->id == device2->id but they are diff device objects (device2 is initialized)

this can be cleaned up when we don't need to create device objects but rather give handle to existing device object

…smem writer (HW CQ) to be owned by Device assign each device to specific host mem channel

…ueueProgramCache

…e number of chips on a card and run multi cq device fixture on GS

…that are upgrading

…eing set and multiply it by 4 for multi CQ (cherry picked from commit 226b550)

…es to try fixing OOM memory error for frequent pipeline on GS

abhullar-tt force-pushed the abhullar/cq-per-device branch 3 times, most recently from cb22adb to 8914d1c Compare December 5, 2023 15:25

abhullar-tt changed the title ~~#2934: WIP Make one CommandQueue and one HW CommandQueue (SysmemWriter) per device~~ #2934: Make one CommandQueue and one HW CommandQueue (SysmemWriter) per device Dec 5, 2023

abhullar-tt marked this pull request as ready for review December 5, 2023 17:36

abhullar-tt requested review from DrJessop, davorchap and TT-billteng as code owners December 5, 2023 17:36

abhullar-tt requested a review from aliuTT December 5, 2023 17:36

abhullar-tt commented Dec 5, 2023

View reviewed changes

tt_metal/detail/tt_metal.hpp Outdated Show resolved Hide resolved

abhullar-tt commented Dec 5, 2023

View reviewed changes

tt_metal/llrt/tt_cluster.cpp Outdated Show resolved Hide resolved

abhullar-tt force-pushed the abhullar/cq-per-device branch from 8914d1c to 8fc6fe3 Compare December 5, 2023 18:23

abhullar-tt force-pushed the abhullar/cq-per-device branch from 8fc6fe3 to 9fc4f0e Compare December 5, 2023 19:04

abhullar-tt requested review from boris-drazic, farbabi and tt-rkim as code owners December 5, 2023 23:53

TT-billteng reviewed Dec 6, 2023

View reviewed changes

tests/tt_metal/tt_metal/unit_tests/compute/matmul/single_core_matmul_compute.cpp Show resolved Hide resolved

TT-billteng reviewed Dec 6, 2023

View reviewed changes

tests/tt_metal/tt_metal/unit_tests_fast_dispatch/command_queue/test_CommandQueue.cpp Show resolved Hide resolved

TT-billteng reviewed Dec 6, 2023

View reviewed changes

tests/tt_metal/tt_metal/unit_tests_fast_dispatch/common/command_queue_fixture.hpp Outdated Show resolved Hide resolved

TT-billteng reviewed Dec 6, 2023

View reviewed changes

tt_metal/impl/dispatch/command_queue_interface.hpp Show resolved Hide resolved

abhullar-tt mentioned this pull request Dec 6, 2023

Enable machines to initialize max number of hugepages per MMIO device #4211

Open

TT-billteng approved these changes Dec 6, 2023

View reviewed changes

DrJessop approved these changes Dec 6, 2023

View reviewed changes

abhullar-tt force-pushed the abhullar/cq-per-device branch from 4d34b98 to f04461f Compare December 7, 2023 01:11

boris-drazic approved these changes Dec 7, 2023

View reviewed changes

abhullar-tt force-pushed the abhullar/cq-per-device branch 2 times, most recently from 9f783b4 to b7da114 Compare December 8, 2023 23:42

abhullar-tt mentioned this pull request Dec 11, 2023

Multiple CQs on one device #4269

Closed

abhullar-tt and others added 8 commits December 20, 2023 15:01

#2934: Replace GLOBAL_CQ with one CommandQueue per device and move sy…

2515b94

…smem writer (HW CQ) to be owned by Device assign each device to specific host mem channel

#2934: Pass device into python calls to Synchronize and ClearCommandQ…

6245721

…ueueProgramCache

#2934 Add check that expects num host channels for a mmio device to b…

55f4b50

…e number of chips on a card and run multi cq device fixture on GS

#2934: Uplift tests to use GetCommandQueue rather than GLOBAL_CQ

149abe8

#4211: (REVERT ME) Target all post commit workflows to test machines …

0b4e13e

…that are upgrading

#4211: Upgrade setup_hugepages.py to centralize number of hugepages b…

bdfda13

…eing set and multiply it by 4 for multi CQ (cherry picked from commit 226b550)

#4211: For Grayskull, set number of hugepages to only number of devic…

c8b3a9f

…es to try fixing OOM memory error for frequent pipeline on GS

#4211: (REVERT ME) Run full models on upgraded HP machines

60cf9a1

abhullar-tt force-pushed the abhullar/cq-per-device branch from 8a58161 to 60cf9a1 Compare December 20, 2023 15:02

#0: Add GetCommandQueue to doc dictionary

8b8c4be

abhullar-tt merged commit 8b8c4be into main Dec 20, 2023
4 checks passed

abhullar-tt had a problem deploying to dev December 20, 2023 15:19 — with GitHub Actions Error

abhullar-tt temporarily deployed to dev December 20, 2023 15:19 — with GitHub Actions Inactive

abhullar-tt had a problem deploying to dev December 20, 2023 15:19 — with GitHub Actions Error

abhullar-tt temporarily deployed to dev December 20, 2023 15:19 — with GitHub Actions Inactive

abhullar-tt had a problem deploying to dev December 20, 2023 15:19 — with GitHub Actions Error

abhullar-tt temporarily deployed to dev December 20, 2023 15:19 — with GitHub Actions Inactive

abhullar-tt temporarily deployed to production December 20, 2023 15:41 — with GitHub Actions Inactive

abhullar-tt had a problem deploying to production December 20, 2023 15:41 — with GitHub Actions Error

abhullar-tt temporarily deployed to production December 20, 2023 15:41 — with GitHub Actions Inactive

abhullar-tt had a problem deploying to production December 20, 2023 15:41 — with GitHub Actions Error

Provide feedback

Saved searches

Use saved searches to filter your results more quickly

#2934: Make one CommandQueue and one HW CommandQueue (SysmemWriter) per device #4077

#2934: Make one CommandQueue and one HW CommandQueue (SysmemWriter) per device #4077

abhullar-tt commented Nov 29, 2023 •

edited

Loading

abhullar-tt commented Dec 5, 2023

abhullar-tt commented Dec 6, 2023

DrJessop Dec 6, 2023

abhullar-tt Dec 6, 2023

DrJessop Dec 6, 2023

abhullar-tt Dec 6, 2023

#2934: Make one CommandQueue and one HW CommandQueue (SysmemWriter) per device #4077

#2934: Make one CommandQueue and one HW CommandQueue (SysmemWriter) per device #4077

Conversation

abhullar-tt commented Nov 29, 2023 • edited Loading

abhullar-tt commented Dec 5, 2023

abhullar-tt commented Dec 6, 2023

DrJessop Dec 6, 2023

Choose a reason for hiding this comment

abhullar-tt Dec 6, 2023

Choose a reason for hiding this comment

DrJessop Dec 6, 2023

Choose a reason for hiding this comment

abhullar-tt Dec 6, 2023

Choose a reason for hiding this comment

abhullar-tt commented Nov 29, 2023 •

edited

Loading