[c10d][xccl2] Port the XcclApi oneCCL abstraction - #6
Open
frost-intel wants to merge 1 commit into
Open
frost-intel wants to merge 1 commit into
frost-intel wants to merge 1 commit into
Conversation
First code commit of the in-tree TorchComms XCCL backend. Ports
`comms/torchcomms/xccl/XcclApi.{hpp,cpp}` to
`torch/csrc/distributed/c10d/xccl2/`, namespace `c10d::xccl2`, guarded by
USE_C10D_XCCL. Nothing references it yet; build wiring lands later in the
stack.
This binds the oneCCL *C* API (`onecclXxx`, <oneapi/ccl.h>) rather than the C++
`ccl::` API used by the existing torch-xpu-ops ProcessGroupXCCL. The C API
mirrors NCCL's shape, which is what lets the backend built on top of it stay
structurally aligned with nccl2's NcclApi.
Deviations from the torchcomms original:
- Streams are `c10::xpu::XPUStream` instead of the torchcomms `xpuStream_t`
alias, so `comms/torchcomms/device/xpu/XpuApi.*` does not need porting.
`xpuStream_t` was already `c10::xpu::XPUStream`, and XPUStream's
`operator sycl::queue*` implicitly converts to the `void* stream` the oneCCL
C API takes.
- `getErrorString` returns `std::string_view` to match nccl2's NcclApi, with a
null check: unlike ncclGetErrorString, oneCCL documents no non-null
guarantee, and `std::string_view(nullptr)` is UB.
- Drops the `<oneapi/ccl.hpp>` include; the C API header is sufficient.
commAbort/commGetAsyncError are kept in the interface but return
onecclNotImplemented. oneCCL declares onecclCommAbort and
onecclCommGetAsyncError with CCL_C_NOT_IMPLEMENTED, which expands to
__attribute__((error(...))) -- referencing them is a *compile-time* error, so
they cannot simply forward. The backend must therefore not depend on
abort-based teardown or async error polling for now.
Verified by compiling the TU with the real torch_xpu flags: 63 exported
symbols with USE_C10D_XCCL, 0 without, and all 26 undefined oneCCL references
resolve against libccl.so.2.
This was referenced Aug 25, 2026
This file contains hidden or bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Sign up for free
to join this conversation on GitHub.
Already have an account?
Sign in to comment
1 participant
Add this suggestion to a batch that can be applied as a single commit.This suggestion is invalid because no changes were made to the code.Suggestions cannot be applied while the pull request is closed.Suggestions cannot be applied while viewing a subset of changes.Only one suggestion per line can be applied in a batch.Add this suggestion to a batch that can be applied as a single commit.Applying suggestions on deleted lines is not supported.You must change the existing code in this line in order to create a valid suggestion.Outdated suggestions cannot be applied.This suggestion has been applied or marked resolved.Suggestions cannot be applied from pending reviews.Suggestions cannot be applied on multi-line comments.Suggestions cannot be applied while the pull request is queued to merge.Suggestion cannot be applied right now. Please check back later.
Stack (bottom → top)
xccl2/01-foundationxccl2/02-apixccl2/03-workxccl2/04-engine-declxccl2/05-engine-implxccl2/06-backend-surfacexccl2/07-build-wiringxccl2/08-testsxccl2/09-integration-testsEach PR is exactly one commit and targets the one below it. Review bottom-up.
First code commit of the in-tree TorchComms XCCL backend. Ports
comms/torchcomms/xccl/XcclApi.{hpp,cpp}totorch/csrc/distributed/c10d/xccl2/, namespacec10d::xccl2, guarded byUSE_C10D_XCCL. Nothing references it yet; build wiring lands later in the
stack.
This binds the oneCCL C API (
onecclXxx, <oneapi/ccl.h>) rather than the C++ccl::API used by the existing torch-xpu-ops ProcessGroupXCCL. The C APImirrors NCCL's shape, which is what lets the backend built on top of it stay
structurally aligned with nccl2's NcclApi.
Deviations from the torchcomms original:
c10::xpu::XPUStreaminstead of the torchcommsxpuStream_talias, so
comms/torchcomms/device/xpu/XpuApi.*does not need porting.xpuStream_twas alreadyc10::xpu::XPUStream, and XPUStream'soperator sycl::queue*implicitly converts to thevoid* streamthe oneCCLC API takes.
getErrorStringreturnsstd::string_viewto match nccl2's NcclApi, with anull check: unlike ncclGetErrorString, oneCCL documents no non-null
guarantee, and
std::string_view(nullptr)is UB.<oneapi/ccl.hpp>include; the C API header is sufficient.commAbort/commGetAsyncError are kept in the interface but return
onecclNotImplemented. oneCCL declares onecclCommAbort and
onecclCommGetAsyncError with CCL_C_NOT_IMPLEMENTED, which expands to
attribute((error(...))) -- referencing them is a compile-time error, so
they cannot simply forward. The backend must therefore not depend on
abort-based teardown or async error polling for now.
Verified by compiling the TU with the real torch_xpu flags: 63 exported
symbols with USE_C10D_XCCL, 0 without, and all 26 undefined oneCCL references
resolve against libccl.so.2.