Move public headers under include/

CMakeLists.txt

@@ -26,11 +26,11 @@ FetchContent_Declare(googletest URL https://github.com/google/googletest/archive
 FetchContent_MakeAvailable(googletest)
 include(GoogleTest)
 
-set(CLANG_FORMAT_SOURCE_DIRS src tests)
+set(CLANG_FORMAT_SOURCE_DIRS include src tests)
 include(${PROJECT_SOURCE_DIR}/cmake/AddClangFormatTargets.cmake)
 
 add_library(mscclpp SHARED)
-target_include_directories(mscclpp PUBLIC ${CMAKE_CURRENT_SOURCE_DIR}/src/include)
+target_include_directories(mscclpp PUBLIC ${CMAKE_CURRENT_SOURCE_DIR}/include PRIVATE ${CMAKE_CURRENT_SOURCE_DIR}/src/include)
 set_target_properties(mscclpp PROPERTIES LINKER_LANGUAGE CXX)
 target_link_libraries(mscclpp PRIVATE MSCCLPP::ibverbs MSCCLPP::numa CUDA::cudart CUDA::cuda_driver)
 if(ENABLE_TRACE)

include/mscclpp/channel.hpp

@@ -1,11 +1,10 @@
 #ifndef MSCCLPP_CHANNEL_HPP_
 #define MSCCLPP_CHANNEL_HPP_
 
-#include "epoch.hpp"
-#include "mscclpp.hpp"
-#include "mscclppfifo.hpp"
-#include "proxy.hpp"
-#include "utils.hpp"
+#include <mscclpp/epoch.hpp>
+#include <mscclpp/core.hpp>
+#include <mscclpp/fifo.hpp>
+#include <mscclpp/proxy.hpp>
 
 namespace mscclpp {
 namespace channel {
@@ -148,7 +147,7 @@ struct DeviceChannel
 
   __forceinline__ __device__ void flush()
   {
-    uint64_t curFifoHead = fifo_.push(ChannelTrigger(mscclppSync, 0, 0, 0, 0, 1, channelId_).value);
+    uint64_t curFifoHead = fifo_.push(ChannelTrigger(TriggerSync, 0, 0, 0, 0, 1, channelId_).value);
     // we need to wait for two conditions to be met to ensure the CPU is done flushing. (1) wait for the tail
     // to go pass by curFifoHead (this is safety net) and (2) wait for the work element value to change to 0.
     while (*(volatile uint64_t*)&fifo_.triggers[curFifoHead % MSCCLPP_PROXY_FIFO_SIZE] != 0 &&

include/mscclpp/core.hpp

@@ -1,12 +1,12 @@
-#ifndef MSCCLPP_HPP_
-#define MSCCLPP_HPP_
+#ifndef MSCCLPP_CORE_HPP_
+#define MSCCLPP_CORE_HPP_
 
 #define MSCCLPP_MAJOR 0
 #define MSCCLPP_MINOR 1
 #define MSCCLPP_PATCH 0
 #define MSCCLPP_VERSION (MSCCLPP_MAJOR * 10000 + MSCCLPP_MINOR * 100 + MSCCLPP_PATCH)
 
-#include "errors.hpp"
+#include <mscclpp/errors.hpp>
 #include <bitset>
 #include <future>
 #include <memory>
@@ -383,4 +383,4 @@ template <> struct hash<mscclpp::TransportFlags>
 };
 } // namespace std
 
-#endif // MSCCLPP_H_
+#endif // MSCCLPP_CORE_HPP_

include/mscclpp/epoch.hpp

@@ -1,7 +1,7 @@
 #ifndef MSCCLPP_EPOCH_HPP_
 #define MSCCLPP_EPOCH_HPP_
 
-#include "mscclpp.hpp"
+#include <mscclpp/core.hpp>
 
 namespace mscclpp {
 

include/mscclpp/errors.hpp

@@ -50,5 +50,7 @@ public:
   IbError(std::string message, int errorCode);
   virtual ~IbError() = default;
 };
+
 };     // namespace mscclpp
-#endif // MSCCLPP_ERRORS_HPP
+
+#endif // MSCCLPP_ERRORS_HPP_

include/mscclpp/fifo.hpp

@@ -1,5 +1,5 @@
-#ifndef MSCCLPPFIFO_HPP_
-#define MSCCLPPFIFO_HPP_
+#ifndef MSCCLPP_FIFO_HPP_
+#define MSCCLPP_FIFO_HPP_
 
 #include <functional>
 #include <memory>
@@ -74,4 +74,4 @@ private:
 
 } // namespace mscclpp
 
-#endif // MSCCLPPFIFO_H_
+#endif // MSCCLPP_FIFO_HPP_

include/mscclpp/proxy.hpp

@@ -1,7 +1,7 @@
 #ifndef MSCCLPP_PROXY_HPP_
 #define MSCCLPP_PROXY_HPP_
 
-#include "mscclppfifo.hpp"
+#include <mscclpp/fifo.hpp>
 
 #include <functional>
 #include <memory>

src/bootstrap/bootstrap.cc

@@ -1,7 +1,7 @@
 #include "bootstrap.h"
 #include "api.h"
 #include "checks.hpp"
-#include "mscclpp.hpp"
+#include <mscclpp/core.hpp>
 #include "utils.h"
 
 #include <algorithm>

src/channel.cc

@@ -1,4 +1,4 @@
-#include "channel.hpp"
+#include <mscclpp/channel.hpp>
 #include "api.h"
 #include "checks.hpp"
 #include "debug.h"

src/communicator.cc

@@ -6,7 +6,7 @@
 #include "communicator.hpp"
 #include "connection.hpp"
 #include "debug.h"
-#include "mscclpp.hpp"
+#include <mscclpp/core.hpp>
 #include "registered_memory.hpp"
 #include "utils.h"
 

src/epoch.cc

@@ -1,4 +1,4 @@
-#include "epoch.hpp"
+#include <mscclpp/epoch.hpp>
 #include "alloc.h"
 #include "api.h"
 #include "checks.hpp"

src/errors.cc

@@ -1,4 +1,4 @@
-#include "errors.hpp"
+#include <mscclpp/errors.hpp>
 #include "api.h"
 
 namespace mscclpp {

src/fifo.cc

@@ -1,7 +1,7 @@
 #include "alloc.h"
 #include "api.h"
 #include "checks.hpp"
-#include "mscclppfifo.hpp"
+#include <mscclpp/fifo.hpp>
 #include <cuda_runtime.h>
 #include <emmintrin.h>
 #include <stdexcept>

src/ib.cc

@@ -11,7 +11,7 @@
 #include "comm.h"
 #include "debug.h"
 #include "ib.hpp"
-#include "mscclpp.hpp"
+#include <mscclpp/core.hpp>
 #include <infiniband/verbs.h>
 #include <string>
 

src/include/basic_proxy_handler.hpp

@@ -2,7 +2,7 @@
 #define MSCCLPP_BASIC_PROXY_SERVICE_HPP_
 
 #include "communicator.hpp"
-#include "mscclpp.hpp"
+#include <mscclpp/core.hpp>
 
 namespace mscclpp {
 

src/include/checks.hpp

@@ -8,7 +8,7 @@
 #define MSCCLPP_CHECKS_HPP_
 
 #include "debug.h"
-#include "errors.hpp"
+#include <mscclpp/errors.hpp>
 
 #include <cuda.h>
 #include <cuda_runtime.h>

src/include/communicator.hpp

@@ -3,8 +3,8 @@
 
 #include "ib.hpp"
 #include "mscclpp.h"
-#include "mscclpp.hpp"
-#include "proxy.hpp"
+#include <mscclpp/core.hpp>
+#include <mscclpp/proxy.hpp>
 #include <memory>
 #include <unordered_map>
 

src/include/connection.hpp

@@ -6,7 +6,7 @@
 
 #include "communicator.hpp"
 #include "ib.hpp"
-#include "mscclpp.hpp"
+#include <mscclpp/core.hpp>
 #include <cuda_runtime.h>
 
 namespace mscclpp {

src/include/registered_memory.hpp

@@ -2,10 +2,10 @@
 #define MSCCLPP_REGISTERED_MEMORY_HPP_
 
 #include "communicator.hpp"
-#include "errors.hpp"
+#include <mscclpp/errors.hpp>
 #include "ib.hpp"
 #include "mscclpp.h"
-#include "mscclpp.hpp"
+#include <mscclpp/core.hpp>
 #include <cuda_runtime.h>
 
 namespace mscclpp {

src/proxy_cpp.cc

@@ -1,6 +1,6 @@
 #include "api.h"
-#include "mscclpp.hpp"
-#include "proxy.hpp"
+#include <mscclpp/core.hpp>
+#include <mscclpp/proxy.hpp>
 #include "utils.h"
 #include "utils.hpp"
 #include <atomic>

test/CMakeLists.txt

@@ -1,6 +1,7 @@
 function(add_test_executable name sources)
     add_executable(${name} ${sources})
     target_link_libraries(${name} mscclpp CUDA::cudart CUDA::cuda_driver)
+    target_include_directories(${name} PRIVATE ${PROJECT_SOURCE_DIR}/src/include)
     if(USE_MPI_FOR_TESTS)
         target_link_libraries(${name} MPI::MPI_CXX)
         target_compile_definitions(${name} PRIVATE MSCCLPP_USE_MPI_FOR_TESTS)

test/allgather_test_cpp.cu

@@ -1,7 +1,6 @@
-#include "mscclpp.h"
-#include "mscclpp.hpp"
+#include <mscclpp/core.hpp>
 
-#include "channel.hpp"
+#include <mscclpp/channel.hpp>
 
 #ifdef MSCCLPP_USE_MPI_FOR_TESTS
 #include "mpi.h"

test/bootstrap_test_cpp.cc

@@ -1,4 +1,4 @@
-#include "mscclpp.hpp"
+#include <mscclpp/core.hpp>
 
 #include <cassert>
 #include <iostream>

test/communicator_test_cpp.cu

@@ -1,5 +1,5 @@
-#include "epoch.hpp"
-#include "mscclpp.hpp"
+#include <mscclpp/epoch.hpp>
+#include <mscclpp/core.hpp>
 
 #include <cassert>
 #include <cuda_runtime.h>

test/ib_test.cc

@@ -2,7 +2,7 @@
 #include "checks.h"
 #include "ib.hpp"
 #include "infiniband/verbs.h"
-#include "mscclpp.hpp"
+#include <mscclpp/core.hpp>
 #include <array>
 #include <string>
 

test/unit/core_tests.cc

@@ -1,6 +1,6 @@
 #include <gtest/gtest.h>
 #include <gmock/gmock.h>
-#include "mscclpp.hpp"
+#include <mscclpp/core.hpp>
 
 class LocalCommunicatorTest : public ::testing::Test {
  protected: