[PR #680] Fix ElasticBuffer GIN assertion failure with MLNX_OFED 24.10 #701

New issue

Open

opened 2026-10-27 21:05:25 +00:00 by bsakaguchi · 0 comments

bsakaguchi commented

2026-10-27 21:05:25 +00:00

Owner

📋 Pull Request Information

Original PR: https://github.com/deepseek-ai/DeepEP/pull/680
Author: @Functionhx
Created: 7/9/2026
Status: 🔄 Open

Base: main ← Head: fix/mlnx-ofed-24.10-compat

📝 Commits (1)

f36b0ec Relax GIN assertion for MLNX_OFED 24.10 compatibility (#628)

📊 Changes

1 file changed (+39 additions, -12 deletions)

View changed files

📝 csrc/kernels/backend/nccl.cu (+39 -12)

📄 Description

What Problem This Solves

Fixes #628: EP_HOST_ASSERT(GIN != NONE) crashes on MLNX_OFED 24.10 because libmlx5 lacks MLX5_1.25 symbols for GIN support.

Fix

Instead of asserting, perform a consensus check across all ranks via ncclAllReduce(ncclMin):

All ranks independently check GIN availability
All-reduce computes minimum across ranks
If ANY rank lacks GIN → all ranks agree to skip it
Graceful fallback to non-GIN path (same as EP_DISABLE_GIN=1)
Debug message wrapped in EP_BUFFER_DEBUG guard (consistent with file convention)
Prints both ginType and railedGinType for accurate diagnostics

This prevents both the crash AND the collective mismatch hang that would occur if ranks independently decided GIN state.

Evidence

cd /home/as/vllm/DeepLearning/DeepEP && git diff main...Functionhx:fix/mlnx-ofed-24.10-compat

Single file changed: csrc/kernels/backend/nccl.cu (+22/-5 lines).

🤖 Generated with Claude Code

_{🔄 This issue represents a GitHub Pull Request. It cannot be merged through Gitea due to API limitations.}

## 📋 Pull Request Information **Original PR:** https://github.com/deepseek-ai/DeepEP/pull/680 **Author:** [@Functionhx](https://github.com/Functionhx) **Created:** 7/9/2026 **Status:** 🔄 Open **Base:** `main` ← **Head:** `fix/mlnx-ofed-24.10-compat` --- ### 📝 Commits (1) - [`f36b0ec`](https://github.com/deepseek-ai/DeepEP/commit/f36b0ecbe95459891d2afacf50795ac13ca11557) Relax GIN assertion for MLNX_OFED 24.10 compatibility (#628) ### 📊 Changes **1 file changed** (+39 additions, -12 deletions) <details> <summary>View changed files</summary> 📝 `csrc/kernels/backend/nccl.cu` (+39 -12) </details> ### 📄 Description ## What Problem This Solves Fixes #628: `EP_HOST_ASSERT(GIN != NONE)` crashes on MLNX_OFED 24.10 because `libmlx5` lacks MLX5_1.25 symbols for GIN support. ## Fix Instead of asserting, perform a consensus check across all ranks via `ncclAllReduce`(`ncclMin`): - All ranks independently check GIN availability - All-reduce computes minimum across ranks - If ANY rank lacks GIN → all ranks agree to skip it - Graceful fallback to non-GIN path (same as `EP_DISABLE_GIN=1`) - Debug message wrapped in `EP_BUFFER_DEBUG` guard (consistent with file convention) - Prints both `ginType` and `railedGinType` for accurate diagnostics This prevents both the crash AND the collective mismatch hang that would occur if ranks independently decided GIN state. ## Evidence ``` cd /home/as/vllm/DeepLearning/DeepEP && git diff main...Functionhx:fix/mlnx-ofed-24.10-compat ``` Single file changed: `csrc/kernels/backend/nccl.cu` (+22/-5 lines). 🤖 Generated with [Claude Code](https://claude.com/claude-code) --- <sub>🔄 This issue represents a GitHub Pull Request. It cannot be merged through Gitea due to API limitations.</sub>

bsakaguchi added the

pull-request

label

2026-10-27 21:05:25 +00:00

bsakaguchi referenced this issue

2026-10-27 21:05:33 +00:00

[PR #701] fix(legacy): avoid undefined 32-bit shift #716