diff --git a/Makefile b/Makefile index 266187ed..4f364341 100644 --- a/Makefile +++ b/Makefile @@ -1,5 +1,6 @@ DEBUG ?= 0 VERBOSE ?= 1 +TRACE ?= 0 ######## CUDA CUDA_HOME ?= /usr/local/cuda @@ -47,6 +48,10 @@ CXXFLAGS := -DCUDA_MAJOR=$(CUDA_MAJOR) -DCUDA_MINOR=$(CUDA_MINOR) -fPIC -fvisi -Wall -Wno-unused-function -Wno-sign-compare -std=c++14 -Wvla \ -I $(CUDA_INC) \ $(CXXFLAGS) + +ifneq ($(TRACE), 0) +CXXFLAGS += -DENABLE_TRACE +endif # Maxrregcount needs to be set accordingly to MSCCLPP_MAX_NTHREADS (otherwise it will cause kernel launch errors) # 512 : 120, 640 : 96, 768 : 80, 1024 : 60 # We would not have to set this if we used __launch_bounds__, but this only works on kernels, not on functions. diff --git a/src/include/alloc.h b/src/include/alloc.h index b0ccd0cd..7b06e659 100644 --- a/src/include/alloc.h +++ b/src/include/alloc.h @@ -46,7 +46,7 @@ mscclppResult_t mscclppCallocDebug(T** ptr, size_t nelem, const char *filefunc, WARN("Failed to malloc %ld bytes", nelem*sizeof(T)); return mscclppSystemError; } - //INFO(MSCCLPP_ALLOC, "%s:%d malloc Size %ld pointer %p", filefunc, line, nelem*sizeof(T), p); + INFO(MSCCLPP_ALLOC, "%s:%d malloc Size %ld pointer %p", filefunc, line, nelem*sizeof(T), p); memset(p, 0, nelem*sizeof(T)); *ptr = (T*)p; return mscclppSuccess; diff --git a/src/init.cc b/src/init.cc index bb9bf10e..a70f3295 100644 --- a/src/init.cc +++ b/src/init.cc @@ -2,6 +2,17 @@ #include "bootstrap.h" #include "core.h" +static uint64_t hashUniqueId(mscclppUniqueId const &id) { + char const *bytes = (char const*)&id; + uint64_t h = 0xdeadbeef; + for(int i=0; i < (int)sizeof(mscclppUniqueId); i++) { + h ^= h >> 32; + h *= 0x8db3db47fa2994ad; + h += bytes[i]; + } + return h; +} + pthread_mutex_t initLock = PTHREAD_MUTEX_INITIALIZER; static bool initialized = false; // static size_t maxLocalSizeBytes = 0; @@ -30,6 +41,6 @@ mscclppResult_t mscclppGetUniqueId(mscclppUniqueId* out) { MSCCLPPCHECK(mscclppInit()); // mscclppCHECK(PtrCheck(out, "GetUniqueId", "out")); mscclppResult_t res = bootstrapGetUniqueId((struct mscclppBootstrapHandle*)out); -// TRACE_CALL("mscclppGetUniqueId(0x%llx)", (unsigned long long)hashUniqueId(*out)); + TRACE_CALL("mscclppGetUniqueId(0x%llx)", (unsigned long long)hashUniqueId(*out)); return res; } \ No newline at end of file diff --git a/src/init_test.cc b/src/init_test.cc index 92c5604f..39e46963 100644 --- a/src/init_test.cc +++ b/src/init_test.cc @@ -9,6 +9,6 @@ int main() printf("mscclppGetUniqueId failed\n"); return -1; } - printf("Succeeded! %d\n", uid); + printf("Succeeded!\n"); return 0; }