Revert disabling of context tracking for Radeon (#1927)

* Revert disabling of context tracking for Radeon

Original commit 6fc228e2
 `Disable context tracking for the current version. (#1839)`

* Add env variable for disabling of context tracking for Radeon

`export NCCL_DISABLE_CONTEXT_TRACKING=1` to force disable of context tracking

* Update docs/how-to/rccl-usage-tips.rst

Fix grammar, thanks @amd-jnovotny

Co-authored-by: Jeffrey Novotny <jnovotny@amd.com>

* Rename NCCL_DISABLE_CONTEXT_TRACKING -> RCCL_DISABLE_CONTEXT_TRACKING

* Revert changes in includes and rename util function

---------

Co-authored-by: Jeffrey Novotny <jnovotny@amd.com>

src/misc/utils.cc

@@ -8,6 +8,8 @@
 #include "core.h"
 
 #include "nvmlwrap.h"
+#include "archinfo.h"
+
 #include <unistd.h>
 #include <stdlib.h>
 
@@ -191,6 +193,18 @@ bool matchIfList(const char* string, int port, struct netIf* ifList, int listSiz
   return false;
 }
 
+RCCL_PARAM(DisableContextTracking, "DISABLE_CONTEXT_TRACKING", 0);
+bool rcclNeedEnableContextTrack(int cuDeviceId) {
+  hipDeviceProp_t devProp;
+  if (rcclParamDisableContextTracking() == 1)
+    return false;
+  if (hipGetDeviceProperties(&devProp, cuDeviceId) != 0)
+    return false;
+  return IsArchMatch(devProp.gcnArchName,"gfx11")
+         || IsArchMatch(devProp.gcnArchName,"gfx12")
+         || IsArchMatch(devProp.gcnArchName,"gfx10");
+}
+
 __thread struct ncclThreadSignal ncclThreadSignalLocalInstance = ncclThreadSignalStaticInitializer();
 
 void* ncclMemoryStack::allocateSpilled(struct ncclMemoryStack* me, size_t size, size_t align) {