From 44eb40da0e9bc80f19772e6dd928990c2a38af36 Mon Sep 17 00:00:00 2001
From: John Bachan <jbachan@nvidia.com>
Date: Mon, 14 Mar 2022 17:09:08 -0700
Subject: [PATCH] Add pthread_detach()'s for threads we never pthread_join().
 Helps reduce diagnostic noise for ThreadSanitizer.

Fixes https://github.com/NVIDIA/nccl/issues/649
---
 src/bootstrap.cc        | 1 +
 src/proxy.cc            | 1 +
 src/transport/net_ib.cc | 1 +
 3 files changed, 3 insertions(+)

diff --git a/src/bootstrap.cc b/src/bootstrap.cc
index db1e70e..daaa8cd 100644
--- a/src/bootstrap.cc
+++ b/src/bootstrap.cc
@@ -165,6 +165,7 @@ ncclResult_t bootstrapCreateRoot(ncclUniqueId* id, bool idFromEnv) {
   memcpy(id, &listenSock->addr, sizeof(union ncclSocketAddress));
   pthread_t thread;
   pthread_create(&thread, NULL, bootstrapRoot, (void*)listenSock);
+  pthread_detach(thread); // will not be pthread_join()'d
   ncclSetThreadName(thread, "NCCL BootstrapR");
   return ncclSuccess;
 }
diff --git a/src/proxy.cc b/src/proxy.cc
index 7d4f811..d6fe309 100644
--- a/src/proxy.cc
+++ b/src/proxy.cc
@@ -1073,6 +1073,7 @@ ncclResult_t ncclProxyInit(struct ncclComm* comm, struct ncclSocket* sock, union
 }
 
 ncclResult_t ncclProxyCreate(struct ncclComm* comm) {
+  // comm->proxyState.thread is pthread_join()'d by commFree() in init.cc
   pthread_create(&comm->proxyState.thread, NULL, ncclProxyService, comm);
   return ncclSuccess;
 }
diff --git a/src/transport/net_ib.cc b/src/transport/net_ib.cc
index 4edff0f..6e8f215 100644
--- a/src/transport/net_ib.cc
+++ b/src/transport/net_ib.cc
@@ -223,6 +223,7 @@ ncclResult_t ncclIbInit(ncclDebugLogger_t logFunction) {
           ncclIbDevs[ncclNIbDevs].mrCache.slots = NULL;
 
           pthread_create(&ncclIbAsyncThread, NULL, ncclIbAsyncThreadMain, context);
+          pthread_detach(ncclIbAsyncThread); // will not be pthread_join()'d
           ncclSetThreadName(ncclIbAsyncThread, "NCCL IbAsync %2d", ncclNIbDevs);
           ncclNIbDevs++;
           nPorts++;