diff --git a/tools/topo_expl/Makefile b/tools/topo_expl/Makefile index dfab36f19c..0e8d708a59 100644 --- a/tools/topo_expl/Makefile +++ b/tools/topo_expl/Makefile @@ -6,7 +6,7 @@ endif HIPCC = $(HIP_PATH)/bin/hipcc EXE = topo_expl -CXXFLAGS = -g -ffunction-sections -fdata-sections -Wl,--gc-sections -fgpu-rdc -Iinclude -Ihipify_rccl/include -Ihipify_rccl/device/include -Ihipify_rccl/graph -I/opt/rocm/include/ -DTOPO_EXPL -DENABLE_TRACE -DENABLE_LL128 -DNVTX_NO_IMPL -DRCCL_EXPOSE_STATIC -lpthread +CXXFLAGS = -g -ffunction-sections -fdata-sections -Wl,--gc-sections -fgpu-rdc -Iinclude -Ihipify_rccl/include -Ihipify_rccl/include/plugin -Ihipify_rccl/device/include -Ihipify_rccl/graph -I/opt/rocm/include/ -DTOPO_EXPL -DENABLE_TRACE -DENABLE_LL128 -DNVTX_NO_IMPL -DRCCL_EXPOSE_STATIC -lpthread files = $(EXE).cpp model.cpp utils.cpp hipify_rccl/graph/topo.cc hipify_rccl/graph/rings.cc hipify_rccl/graph/paths.cc hipify_rccl/graph/trees.cc ../../src/misc/param.cc \ hipify_rccl/graph/search.cc hipify_rccl/graph/connect.cc hipify_rccl/graph/tuning.cc hipify_rccl/graph/xml.cc ../../src/misc/nvmlwrap_stub.cc hipify_rccl/graph/rome_models.cc hipify_rccl/graph/archinfo.cc \ @@ -29,6 +29,7 @@ hipify: cp -a ../../src/collectives.cc hipify_rccl/ cp -a ../../src/misc/archinfo.cc hipify_rccl/graph/ hipify-perl -inplace -quiet-warnings hipify_rccl/include/*.h + hipify-perl -inplace -quiet-warnings hipify_rccl/include/plugin/*.h hipify-perl -inplace -quiet-warnings hipify_rccl/device/include/*.h sed -i "s/template/template/g" "hipify_rccl/device/include/common.h" sed -i "s/\\(struct RunWorkBatch]*\\)>*/\\1, COLL_UNROLL>/" "hipify_rccl/device/include/common.h" @@ -38,4 +39,4 @@ hipify: clean: rm -rf hipify_rccl - rm -f *.o $(EXE) \ No newline at end of file + rm -f *.o $(EXE) diff --git a/tools/topo_expl/model.cpp b/tools/topo_expl/model.cpp index 12f22be219..d3ed279a2d 100644 --- a/tools/topo_expl/model.cpp +++ b/tools/topo_expl/model.cpp @@ -191,7 +191,7 @@ struct setupReq { int remoteRank; int shared; int netDev; - int useGdr; + enum ncclTopoGdrMode useGdr; int channelId; int connIndex; }; @@ -263,7 +263,8 @@ ncclResult_t collNetCanConnect(int* ret, struct ncclComm* comm, struct ncclTopo } ncclResult_t collNetSendSetup(struct ncclComm* comm, struct ncclTopoGraph* graph, struct ncclPeerInfo* myInfo, struct ncclPeerInfo* peerInfo, struct ncclConnect* connectInfo, struct ncclConnector* send, int channelId, int connIndex) { - int netDev, useGdr = 0, proxy; + int netDev, proxy; + enum ncclTopoGdrMode useGdr = ncclTopoGdrModeDisable; NCCLCHECK(ncclTopoGetNetDev(comm, myInfo->rank, graph, channelId, peerInfo->rank, NULL, &netDev, &proxy)); NCCLCHECK(ncclTopoCheckGdr(comm->topo, myInfo->rank, netDev, 1, &useGdr)); @@ -273,7 +274,8 @@ ncclResult_t collNetSendSetup(struct ncclComm* comm, struct ncclTopoGraph* graph } ncclResult_t collNetRecvSetup(struct ncclComm* comm, struct ncclTopoGraph* graph, struct ncclPeerInfo* myInfo, struct ncclPeerInfo* peerInfo, struct ncclConnect* connectInfo, struct ncclConnector* recv, int channelId, int connIndex) { - int netDev, useGdr = 0, proxy; + int netDev, proxy; + enum ncclTopoGdrMode useGdr = ncclTopoGdrModeDisable; NCCLCHECK(ncclTopoGetNetDev(comm, myInfo->rank, graph, channelId, peerInfo->rank, NULL, &netDev, &proxy)); NCCLCHECK(ncclTopoCheckGdr(comm->topo, myInfo->rank, netDev, 0, &useGdr));