Skip to content

[Feature Request]: MACA Support for Mooncake EP/PG #1965

Closed

Description

@Dayuxiaoshui

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

  1. Directory naming: maca-ep/ vs mooncake-ep-maca/?
  2. PG setup.py: separate maca-pg/ directory, or inject via env vars?
  3. -include cuda_alike.h via compiler flags (zero source changes) vs explicit #include in headers (one-line change, aligns with transfer-engine convention)?
  4. engine.so is empty; libtransfer_engine.so + libasio.so need proper CMake install rules.

Acceptance Criteria

  • maca.h contains fabric memory stubs
  • mooncake-ep/ and mooncake-pg/ source files remain untouched
  • MACA EP compiles & loads (kernel throws runtime_error at call time — expected)
  • MACA PG compiles & loads (functional)
  • Zero #ifdef USE_MACA in any existing source file

Before submitting a new issue...

  • Make sure you already searched for relevant issues and read the documentation

Activity

  1. Dayuxiaoshui commented on Apr 24, 2026

    @Dayuxiaoshui
    CollaboratorAuthor
  2. stmatengss commented on Apr 29, 2026

    @stmatengss
    Collaborator
  3. UNIDY2002 commented on May 18, 2026

    @UNIDY2002
    Collaborator

    @Dayuxiaoshui Thank you for your nice proposal.

    The overall roadmap is valid, and here are my thoughts to your open questions:

    1. Directory naming: maca-ep/ vs mooncake-ep-maca/? I think the layout proposed in [RFC]: Mooncake EP Multi-Vendor GPU Adaptation Design #1975 is by far the cleanest solution.
    2. PG setup.py: separate maca-pg/ directory, or inject via env vars? I'd prefer injecting via env vars or flags, to avoid duplicating shared logic.
    3. I think adding explicit include directives is better.
    4. engine.so is empty: this is a building system issue.

    One followup question: does #1975 supersede this issue in terms of the Mooncake EP part?

Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment

Metadata

Metadata

Assignees

Labels

No labels
No labels

Type

No type

Projects

No projects

    Milestone

    No milestone

    Relationships

    None yet

    Development

    No branches or pull requests

    Issue actions