Repository navigation
Description
Describe your feature request
Summary
Mooncake EP/PG Python extensions are currently CUDA-only. We need MACA support via cu-bridge without modifying existing source files.
Key Finding
MUSA and HIP do NOT have EP/PG support either — zero references to USE_MUSA/USE_HIP in mooncake-ep/ or mooncake-pg/. This adaptation would be breaking new ground.
Blockers
| Issue | Why |
|---|---|
| Fabric constants missing | CU_MEM_HANDLE_TYPE_FABRIC, CU_DEVICE_ATTRIBUTE_HANDLE_TYPE_FABRIC_SUPPORTED undefined in maca.h and cu-bridge |
maca.h / cu-bridge conflict |
#define cuDeviceGetAttribute mcDeviceGetAttribute collides with cu-bridge's wcuDeviceGetAttribute |
| EP kernel uses PTX asm | cooperative_groups, cuda::atomic, __byte_perm, clock64() cannot be translated by any header bridge |
| Setup.py is standalone | CUDAExtension does not inherit CMake config; MACA paths/libs must be injected separately |
Proposed Design
Principle: Zero modifications to existing .cpp/.cu/.h files.
1. gpu_vendor/maca.h — Add fabric stubs
static const mcMemAllocationHandleType CU_MEM_HANDLE_TYPE_FABRIC = mcMemHandleTypeNone;
static const mcDrvDeviceAttribute_t CU_DEVICE_ATTRIBUTE_HANDLE_TYPE_FABRIC_SUPPORTED =
static_cast<mcDrvDeviceAttribute_t>(0x7FFFFFFF);
#define CUmemFabricHandle void*Remove #define cuDeviceGetAttribute mcDeviceGetAttribute (let cu-bridge handle it).
2. PG — Header mapping only, no source changes
PG uses standard CUDA Runtime APIs. After maca.h is fixed, it should compile as-is.
3. EP Kernel — File-level separation (Ascend-style)
Do not add #ifdef USE_MACA to mooncake_ep_kernel.cu.
Create independent directory:
maca-ep/ # or mooncake-ep-maca/
├── setup.py
└── src/kernel/
└── mooncake_ep_kernel_maca.cpp # stub: throw runtime_error
maca-ep/setup.py references unchanged files from ../mooncake-ep/src/ and swaps only the kernel source.
4. Build system selection
BuildEpExt.cmake chooses directory based on USE_MACA:
if(USE_MACA)
set(EP_SOURCE_DIR "${CMAKE_CURRENT_SOURCE_DIR}/maca-ep")
else()
set(EP_SOURCE_DIR "${CMAKE_CURRENT_SOURCE_DIR}/mooncake-ep")
endif()Open Questions
- Directory naming:
maca-ep/vsmooncake-ep-maca/? - PG setup.py: separate
maca-pg/directory, or inject via env vars? -include cuda_alike.hvia compiler flags (zero source changes) vs explicit#includein headers (one-line change, aligns with transfer-engine convention)?engine.sois empty;libtransfer_engine.so+libasio.soneed proper CMake install rules.
Acceptance Criteria
-
maca.hcontains fabric memory stubs -
mooncake-ep/andmooncake-pg/source files remain untouched - MACA EP compiles & loads (kernel throws
runtime_errorat call time — expected) - MACA PG compiles & loads (functional)
- Zero
#ifdef USE_MACAin any existing source file
Before submitting a new issue...
- Make sure you already searched for relevant issues and read the documentation
Activity
Dayuxiaoshui commentedon Apr 24, 2026
cc @stmatengss
stmatengss commentedon Apr 29, 2026
@UNIDY2002 PTAL
[TransferEngine][Integration] feat: add MACA/MetaX GPU support and fi…
[TransferEngine][Integration] feat: add MACA/MetaX GPU support and fi…
UNIDY2002 commentedon May 18, 2026
@Dayuxiaoshui Thank you for your nice proposal.
The overall roadmap is valid, and here are my thoughts to your open questions:
maca-ep/vsmooncake-ep-maca/? I think the layout proposed in [RFC]: Mooncake EP Multi-Vendor GPU Adaptation Design #1975 is by far the cleanest solution.maca-pg/directory, or inject via env vars? I'd prefer injecting via env vars or flags, to avoid duplicating shared logic.engine.sois empty: this is a building system issue.One followup question: does #1975 supersede this issue in terms of the Mooncake EP part?
[TransferEngine][Integration] feat: add MACA/MetaX GPU support and fi…