Skip to content

Commit 4228e79

Browse files
yangchaossyangchaoss
committed
test(adaptor): make IPC memory handle tests backend-agnostic
1 parent 02fbd40 commit 4228e79

3 files changed

Lines changed: 114 additions & 56 deletions

File tree

‎test/unittest/adaptor/Makefile‎

Lines changed: 5 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -9,6 +9,8 @@ UNIT_TARGET := $(BINDIR)/adaptor_unit_tests
99
MPI_SRCS := $(wildcard coll_*.cpp)
1010
MPI_OBJS := $(MPI_SRCS:%.cpp=$(OBJDIR)/%.o)
1111
MPI_TARGET := $(BINDIR)/adaptor_mpi_tests
12+
MPIRUN ?= mpirun
13+
MPI_NP ?= 2
1214

1315
.PHONY: all clean run-unit run-mpi run
1416

@@ -35,12 +37,12 @@ $(MPI_TARGET): $(MPI_OBJS)
3537
$(OBJDIR)/test_%.o: test_%.cpp
3638
@mkdir -p $(OBJDIR)
3739
@echo "Compiling $<"
38-
@$(CXX) $< -o $@ -c $(CXXFLAGS) $(ADAPTOR_DEFINE) $(FLAGCX_INCLUDES) $(DEVICE_INCLUDE) $(CCL_INCLUDE) $(GTEST_INCLUDE) $(FIXTURE_INCLUDE)
40+
@$(CXX) $< -o $@ -c $(CXXFLAGS) $(FLAGCX_INCLUDES) $(GTEST_INCLUDE) $(FIXTURE_INCLUDE)
3941

4042
$(OBJDIR)/coll_%.o: coll_%.cpp
4143
@mkdir -p $(OBJDIR)
4244
@echo "Compiling $<"
43-
@$(CXX) $< -o $@ -c $(CXXFLAGS) $(ADAPTOR_DEFINE) $(FLAGCX_INCLUDES) $(GTEST_INCLUDE) $(FIXTURE_INCLUDE) $(MPI_INCLUDE) $(DEVICE_INCLUDE) $(CCL_INCLUDE)
45+
@$(CXX) $< -o $@ -c $(CXXFLAGS) $(FLAGCX_INCLUDES) $(GTEST_INCLUDE) $(FIXTURE_INCLUDE) $(MPI_INCLUDE)
4446

4547
clean:
4648
@rm -rf $(BUILDDIR)
@@ -49,6 +51,6 @@ run-unit: $(UNIT_TARGET)
4951
@$(UNIT_TARGET)
5052

5153
run-mpi: $(MPI_TARGET)
52-
@mpirun --allow-run-as-root -np 8 $(MPI_TARGET)
54+
@$(MPIRUN) --allow-run-as-root -np $(MPI_NP) $(MPI_TARGET)
5355

5456
run: run-unit run-mpi

‎test/unittest/adaptor/coll_ipc_mem_handle.cpp‎

Lines changed: 64 additions & 22 deletions
Original file line numberDiff line numberDiff line change
@@ -1,6 +1,9 @@
11
/*************************************************************************
22
* Copyright (c) 2026. All Rights Reserved.
3-
* Cross-process IPC memory handle test (requires MPI and KunlunXin XPU).
3+
* Cross-process IPC memory handle test.
4+
*
5+
* Run exactly two MPI ranks on the same host. Device IPC handles are
6+
* host-local and cannot be transferred between different nodes.
47
************************************************************************/
58

69
#include <gtest/gtest.h>
@@ -32,16 +35,25 @@ namespace {
3235
} \
3336
} while (0)
3437

38+
#define ASSERT_MPI_TRUE(condition) \
39+
do { \
40+
if (!(condition)) { \
41+
ADD_FAILURE() << "MPI assertion failed: " << #condition; \
42+
MPI_Abort(MPI_COMM_WORLD, 1); \
43+
return; \
44+
} \
45+
} while (0)
46+
3547
class IpcMemHandleMpiTest : public ::testing::Test {
3648
protected:
3749
void SetUp() override {
3850
flagcxDeviceHandleInit(&devHandle);
39-
ASSERT_NE(devHandle, nullptr);
51+
ASSERT_MPI_TRUE(devHandle != nullptr);
4052

4153
int deviceCount = 0;
4254
ASSERT_FLAGCX_SUCCESS(devHandle->getDeviceCount(&deviceCount));
4355
if (deviceCount <= 0) {
44-
ADD_FAILURE() << "No visible XPU device";
56+
ADD_FAILURE() << "No visible device";
4557
MPI_Abort(MPI_COMM_WORLD, 1);
4658
return;
4759
}
@@ -70,35 +82,70 @@ TEST_F(IpcMemHandleMpiTest, CrossProcessLifecycle) {
7082
constexpr size_t bufferSize = 4096;
7183
constexpr int expectedValue = 0x12345678;
7284

85+
int localApisAvailable =
86+
devHandle->ipcMemHandleCreate != nullptr &&
87+
devHandle->ipcMemHandleGet != nullptr &&
88+
devHandle->ipcMemHandleOpen != nullptr &&
89+
devHandle->ipcMemHandleClose != nullptr &&
90+
devHandle->ipcMemHandleFree != nullptr;
91+
int allApisAvailable = 0;
92+
ASSERT_MPI_SUCCESS(MPI_Allreduce(&localApisAvailable, &allApisAvailable, 1,
93+
MPI_INT, MPI_MIN, MPI_COMM_WORLD));
94+
if (!allApisAvailable) {
95+
GTEST_SKIP() << "IPC memory handle APIs are not available";
96+
}
97+
98+
// Every rank creates its receive/export storage before communication. This
99+
// also provides a coordinated runtime capability check for stub backends.
100+
flagcxIpcMemHandle_t handle = nullptr;
101+
size_t localHandleSize = 0;
102+
flagcxResult_t createResult =
103+
devHandle->ipcMemHandleCreate(&handle, &localHandleSize);
104+
if (createResult != flagcxSuccess &&
105+
createResult != flagcxNotSupported) {
106+
ADD_FAILURE() << "ipcMemHandleCreate returned "
107+
<< static_cast<int>(createResult);
108+
MPI_Abort(MPI_COMM_WORLD, static_cast<int>(createResult));
109+
return;
110+
}
111+
int localSupported = createResult != flagcxNotSupported;
112+
int allSupported = 0;
113+
ASSERT_MPI_SUCCESS(MPI_Allreduce(&localSupported, &allSupported, 1, MPI_INT,
114+
MPI_MIN, MPI_COMM_WORLD));
115+
if (!allSupported) {
116+
if (createResult == flagcxSuccess && handle != nullptr) {
117+
ASSERT_FLAGCX_SUCCESS(devHandle->ipcMemHandleFree(handle));
118+
}
119+
GTEST_SKIP() << "IPC memory handles are not supported";
120+
}
121+
ASSERT_FLAGCX_SUCCESS(createResult);
122+
ASSERT_MPI_TRUE(handle != nullptr);
123+
ASSERT_MPI_TRUE(localHandleSize > 0);
124+
73125
if (rank == 0) {
74126
void *devPtr = nullptr;
75127
ASSERT_FLAGCX_SUCCESS(devHandle->deviceMalloc(
76128
&devPtr, bufferSize, flagcxMemDevice, nullptr));
77-
ASSERT_NE(devPtr, nullptr);
129+
ASSERT_MPI_TRUE(devPtr != nullptr);
78130

79131
int hostValue = expectedValue;
80132
ASSERT_FLAGCX_SUCCESS(devHandle->deviceMemcpy(
81133
devPtr, &hostValue, sizeof(hostValue), flagcxMemcpyHostToDevice,
82134
nullptr));
83135
ASSERT_FLAGCX_SUCCESS(devHandle->streamSynchronize(nullptr));
84136

85-
flagcxIpcMemHandle_t handle = nullptr;
86-
size_t handleSize = 0;
87-
ASSERT_FLAGCX_SUCCESS(
88-
devHandle->ipcMemHandleCreate(&handle, &handleSize));
89-
ASSERT_NE(handle, nullptr);
90-
ASSERT_GT(handleSize, static_cast<size_t>(0));
91137
ASSERT_FLAGCX_SUCCESS(devHandle->ipcMemHandleGet(handle, devPtr));
92138

93-
ASSERT_MPI_SUCCESS(MPI_Send(&handleSize, sizeof(handleSize), MPI_BYTE, 1,
94-
0, MPI_COMM_WORLD));
95-
ASSERT_MPI_SUCCESS(MPI_Send(handle, static_cast<int>(handleSize), MPI_BYTE,
96-
1, 1, MPI_COMM_WORLD));
139+
ASSERT_MPI_SUCCESS(MPI_Send(&localHandleSize, sizeof(localHandleSize),
140+
MPI_BYTE, 1, 0, MPI_COMM_WORLD));
141+
ASSERT_MPI_SUCCESS(
142+
MPI_Send(handle, static_cast<int>(localHandleSize), MPI_BYTE, 1, 1,
143+
MPI_COMM_WORLD));
97144

98145
int acknowledgement = 0;
99146
ASSERT_MPI_SUCCESS(MPI_Recv(&acknowledgement, 1, MPI_INT, 1, 2,
100147
MPI_COMM_WORLD, MPI_STATUS_IGNORE));
101-
ASSERT_EQ(acknowledgement, 1);
148+
ASSERT_MPI_TRUE(acknowledgement == 1);
102149

103150
ASSERT_FLAGCX_SUCCESS(devHandle->ipcMemHandleFree(handle));
104151
ASSERT_FLAGCX_SUCCESS(
@@ -109,11 +156,6 @@ TEST_F(IpcMemHandleMpiTest, CrossProcessLifecycle) {
109156
MPI_BYTE, 0, 0, MPI_COMM_WORLD,
110157
MPI_STATUS_IGNORE));
111158

112-
flagcxIpcMemHandle_t handle = nullptr;
113-
size_t localHandleSize = 0;
114-
ASSERT_FLAGCX_SUCCESS(
115-
devHandle->ipcMemHandleCreate(&handle, &localHandleSize));
116-
ASSERT_NE(handle, nullptr);
117159
if (localHandleSize != receivedHandleSize) {
118160
ADD_FAILURE() << "IPC handle size mismatch: local=" << localHandleSize
119161
<< ", remote=" << receivedHandleSize;
@@ -127,14 +169,14 @@ TEST_F(IpcMemHandleMpiTest, CrossProcessLifecycle) {
127169

128170
void *mappedPtr = nullptr;
129171
ASSERT_FLAGCX_SUCCESS(devHandle->ipcMemHandleOpen(handle, &mappedPtr));
130-
ASSERT_NE(mappedPtr, nullptr);
172+
ASSERT_MPI_TRUE(mappedPtr != nullptr);
131173

132174
int receivedValue = 0;
133175
ASSERT_FLAGCX_SUCCESS(devHandle->deviceMemcpy(
134176
&receivedValue, mappedPtr, sizeof(receivedValue),
135177
flagcxMemcpyDeviceToHost, nullptr));
136178
ASSERT_FLAGCX_SUCCESS(devHandle->streamSynchronize(nullptr));
137-
EXPECT_EQ(receivedValue, expectedValue);
179+
ASSERT_MPI_TRUE(receivedValue == expectedValue);
138180

139181
ASSERT_FLAGCX_SUCCESS(devHandle->ipcMemHandleClose(mappedPtr));
140182
ASSERT_FLAGCX_SUCCESS(devHandle->ipcMemHandleFree(handle));

‎test/unittest/adaptor/test_device_adaptor.cpp‎

Lines changed: 45 additions & 31 deletions
Original file line numberDiff line numberDiff line change
@@ -3,7 +3,6 @@
33
* Single device adaptor test - no multi-GPU or MPI required
44
************************************************************************/
55

6-
#include <cstdlib>
76
#include <cstring>
87
#include <gtest/gtest.h>
98
#include <iostream>
@@ -12,10 +11,6 @@
1211
#include "flagcx.h"
1312
#include "topo.h"
1413

15-
#ifdef USE_KUNLUNXIN_ADAPTOR
16-
#include "kunlunxin_adaptor.h"
17-
#endif
18-
1914
class DeviceAdaptorTest : public ::testing::Test {
2015
protected:
2116
void SetUp() override {
@@ -446,36 +441,41 @@ TEST_F(DeviceAdaptorTest, StreamCopyAndFree) {
446441
// Clean up the original
447442
devHandle->streamDestroy(tempStream);
448443
}
449-
#ifdef USE_KUNLUNXIN_ADAPTOR
450-
// Test: ipcMemHandleCreate allocates a wrapper and reports its serialized size.
451-
// Get/Open/Close/Free are intentionally not called in this test.
444+
// Test ipcMemHandleCreate through the public opaque-handle contract.
445+
// ipcMemHandleFree is used only to release resources created by this test.
452446
TEST_F(DeviceAdaptorTest, IpcMemHandleCreate) {
453-
ASSERT_NE(devHandle->ipcMemHandleCreate, nullptr);
447+
if (devHandle->ipcMemHandleCreate == nullptr ||
448+
devHandle->ipcMemHandleFree == nullptr) {
449+
GTEST_SKIP() << "IPC memory handle APIs are not available";
450+
}
454451

455452
flagcxIpcMemHandle_t handle = nullptr;
456453
size_t handleSize = 0;
457-
EXPECT_EQ(devHandle->ipcMemHandleCreate(&handle, &handleSize),
458-
flagcxSuccess);
459-
EXPECT_NE(handle, nullptr);
454+
flagcxResult_t result =
455+
devHandle->ipcMemHandleCreate(&handle, &handleSize);
456+
if (result == flagcxNotSupported) {
457+
GTEST_SKIP() << "IPC memory handles are not supported";
458+
}
459+
ASSERT_EQ(result, flagcxSuccess);
460+
ASSERT_NE(handle, nullptr);
460461
EXPECT_GT(handleSize, static_cast<size_t>(0));
462+
EXPECT_EQ(devHandle->ipcMemHandleFree(handle), flagcxSuccess);
461463

462464
flagcxIpcMemHandle_t handleWithoutSize = nullptr;
463-
EXPECT_EQ(devHandle->ipcMemHandleCreate(&handleWithoutSize, nullptr),
465+
ASSERT_EQ(devHandle->ipcMemHandleCreate(&handleWithoutSize, nullptr),
464466
flagcxSuccess);
465-
EXPECT_NE(handleWithoutSize, nullptr);
466-
467-
EXPECT_EQ(devHandle->ipcMemHandleCreate(nullptr, &handleSize),
468-
flagcxInvalidArgument);
469-
470-
// Direct cleanup keeps this test independent of ipcMemHandleFree.
471-
std::free(handle);
472-
std::free(handleWithoutSize);
467+
ASSERT_NE(handleWithoutSize, nullptr);
468+
EXPECT_EQ(devHandle->ipcMemHandleFree(handleWithoutSize), flagcxSuccess);
473469
}
474470

475-
// Test: ipcMemHandleGet exports a handle from a live device allocation.
476-
// The wrapper is allocated directly so Create/Open/Close/Free are not involved.
471+
// Test ipcMemHandleGet with a handle created through the public API.
472+
// Create/Free are fixture operations; Get remains the assertion target.
477473
TEST_F(DeviceAdaptorTest, IpcMemHandleGet) {
478-
ASSERT_NE(devHandle->ipcMemHandleGet, nullptr);
474+
if (devHandle->ipcMemHandleCreate == nullptr ||
475+
devHandle->ipcMemHandleGet == nullptr ||
476+
devHandle->ipcMemHandleFree == nullptr) {
477+
GTEST_SKIP() << "IPC memory handle APIs are not available";
478+
}
479479

480480
constexpr size_t bufferSize = 4096;
481481
void *devPtr = nullptr;
@@ -485,28 +485,42 @@ TEST_F(DeviceAdaptorTest, IpcMemHandleGet) {
485485
ASSERT_NE(devPtr, nullptr);
486486

487487
flagcxIpcMemHandle_t handle = nullptr;
488-
flagcxCalloc(&handle, 1);
489-
ASSERT_NE(handle, nullptr);
488+
flagcxResult_t result = devHandle->ipcMemHandleCreate(&handle, nullptr);
489+
if (result == flagcxNotSupported) {
490+
EXPECT_EQ(devHandle->deviceFree(devPtr, flagcxMemDevice, stream),
491+
flagcxSuccess);
492+
GTEST_SKIP() << "IPC memory handles are not supported";
493+
}
494+
if (result != flagcxSuccess || handle == nullptr) {
495+
EXPECT_EQ(devHandle->deviceFree(devPtr, flagcxMemDevice, stream),
496+
flagcxSuccess);
497+
FAIL() << "ipcMemHandleCreate returned " << static_cast<int>(result);
498+
}
490499

491500
EXPECT_EQ(devHandle->ipcMemHandleGet(handle, devPtr), flagcxSuccess);
492501
EXPECT_EQ(devHandle->ipcMemHandleGet(nullptr, devPtr),
493502
flagcxInvalidArgument);
494503
EXPECT_EQ(devHandle->ipcMemHandleGet(handle, nullptr),
495504
flagcxInvalidArgument);
496505

497-
std::free(handle);
506+
EXPECT_EQ(devHandle->ipcMemHandleFree(handle), flagcxSuccess);
498507
EXPECT_EQ(devHandle->deviceFree(devPtr, flagcxMemDevice, stream),
499508
flagcxSuccess);
500509
}
501510

502511
// Test: ipcMemHandleClose rejects a null mapped pointer.
503512
// Successful Close is covered by the MPI lifecycle test.
504513
TEST_F(DeviceAdaptorTest, IpcMemHandleClose) {
505-
ASSERT_NE(devHandle->ipcMemHandleClose, nullptr);
506-
EXPECT_EQ(devHandle->ipcMemHandleClose(nullptr), flagcxInvalidArgument);
507-
}
514+
if (devHandle->ipcMemHandleClose == nullptr) {
515+
GTEST_SKIP() << "ipcMemHandleClose is not available";
516+
}
508517

509-
#endif
518+
flagcxResult_t result = devHandle->ipcMemHandleClose(nullptr);
519+
if (result == flagcxNotSupported) {
520+
GTEST_SKIP() << "IPC memory handles are not supported";
521+
}
522+
EXPECT_EQ(result, flagcxInvalidArgument);
523+
}
510524

511525
int main(int argc, char **argv) {
512526
::testing::InitGoogleTest(&argc, argv);

0 commit comments

Comments
 (0)