AI/rocm-systems
2
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

Limit network profiling support to simple protocol and avoid overflow

Browse Source 
[ROCm/rccl commit: 95b8f70d15]
This commit is contained in:
Wenkai Du
2020-06-15 20:51:36 +00:00
parent 3389838def
commit 5d93f83ae3
2 changed files with 23 additions and 13 deletions
Download Patch File Download Diff File Expand all files Collapse all files
projects/rccl/src/include/devcomm.h
+1 -1
View File
@@ -222,7 +222,7 @@ struct ncclChannel {
uint32_t* sync;
#ifdef ENABLE_PROFILING
struct timeval tvs;
int sizes;
uint64_t sizes;
int active_req;
uint64_t send_byte;
uint64_t recv_byte;
projects/rccl/src/transport/net.cc
+22 -12
View File
@@ -340,12 +340,17 @@ ncclResult_t netSendProxy(struct ncclProxyArgs* args) {
NCCLCHECK(ncclNetTest(args->requests[buffSlot], &done, NULL));
if (done) {
#ifdef ENABLE_PROFILING
args->channel->active_req --;
if (args->channel->active_req == 0) {
struct timeval tv;
gettimeofday(&tv, NULL);
args->channel->bw_cumulative += (float)args->channel->sizes/((tv.tv_sec - args->channel->tvs.tv_sec)*1000*1000 + tv.tv_usec - args->channel->tvs.tv_usec)/1000.0;
args->channel->bw_count ++;
if (args->protocol == NCCL_PROTO_SIMPLE) {
args->channel->active_req --;
if (args->channel->active_req == 0) {
struct timeval tv;
gettimeofday(&tv, NULL);
float delta = (tv.tv_sec - args->channel->tvs.tv_sec)*1E6 + tv.tv_usec - args->channel->tvs.tv_usec;
if (delta) {
args->channel->bw_cumulative += (float)args->channel->sizes/delta/1E3;
args->channel->bw_count ++;
}
}
}
#endif
args->head += args->sliceSteps;
@@ -390,11 +395,13 @@ ncclResult_t netRecvProxy(struct ncclProxyArgs* args) {
NCCLCHECK(ncclNetIrecv(resources->netRecvComm, localBuff+buffSlot*stepSize, sliceSize, mhandle, args->requests+buffSlot));
if (args->requests[buffSlot] != NULL) {
#ifdef ENABLE_PROFILING
if (args->channel->active_req == 0) {
gettimeofday(&args->channel->tvs, NULL);
args->channel->sizes = 0;
if (args->protocol == NCCL_PROTO_SIMPLE) {
if (args->channel->active_req == 0) {
gettimeofday(&args->channel->tvs, NULL);
args->channel->sizes = 0;
}
args->channel->active_req ++;
}
args->channel->active_req ++;
#endif
args->tail += args->sliceSteps;
args->idle = 0;
@@ -414,8 +421,11 @@ ncclResult_t netRecvProxy(struct ncclProxyArgs* args) {
if (args->channel->active_req == 0) {
struct timeval tv;
gettimeofday(&tv, NULL);
args->channel->bw_cumulative += (float)args->channel->sizes/((tv.tv_sec - args->channel->tvs.tv_sec)*1000*1000 + tv.tv_usec - args->channel->tvs.tv_usec)/1000.0;
args->channel->bw_count ++;
float delta = (tv.tv_sec - args->channel->tvs.tv_sec)*1E6 + tv.tv_usec - args->channel->tvs.tv_usec;
if (delta) {
args->channel->bw_cumulative += (float)args->channel->sizes/delta/1E3;
args->channel->bw_count ++;
}
}
#endif
if (resources->useGdr) NCCLCHECK(ncclNetFlush(resources->netRecvComm, localBuff+buffSlot*stepSize, size, mhandle));