diff --git a/projects/rocshmem/CMakeLists.txt b/projects/rocshmem/CMakeLists.txt index ee7f42fb12..8603737308 100644 --- a/projects/rocshmem/CMakeLists.txt +++ b/projects/rocshmem/CMakeLists.txt @@ -60,7 +60,6 @@ option(USE_IPC "Enable IPC support (using HIP)" OFF) option(USE_THREADS "Enable workgroup threads to share network queues" OFF) option(USE_WF_COAL "Enable wavefront message coalescing" OFF) option(USE_COHERENT_HEAP "Enable support for coherent systems" OFF) -option(USE_CACHED_HEAP "Enable support for cached systems" OFF) option(USE_MANAGED_HEAP "Enable managed memory" OFF) option(USE_HOST_HEAP "Enable host memory using malloc/free" OFF) option(USE_HIP_HOST_HEAP "Enable host memory using hip api" OFF) diff --git a/projects/rocshmem/README.md b/projects/rocshmem/README.md index c974fd7052..ba671ff51a 100644 --- a/projects/rocshmem/README.md +++ b/projects/rocshmem/README.md @@ -45,7 +45,7 @@ ROC_SHMEM base requirements: * May work with other versions, but not tested * AMD GFX9 GPUs (e.g.: MI25, Vega 56, Vega 64, MI50, MI60, MI100, Radeon VII) * AMD MI200 GPUs: To enable the support on MI200, please configure the library - with USE_CACHED_HEAP + with USE_COHERENT_HEAP * ROCm-aware MPI as described in [Building the Dependencies](#building-the-dependencies) * InfiniBand adaptor compatable with ROCm RDMA technology diff --git a/projects/rocshmem/cmake/config.h.in b/projects/rocshmem/cmake/config.h.in index 6ec07c2853..f409c10ba3 100644 --- a/projects/rocshmem/cmake/config.h.in +++ b/projects/rocshmem/cmake/config.h.in @@ -7,10 +7,9 @@ #cmakedefine USE_SHARED_CTX #cmakedefine USE_WF_COAL #cmakedefine USE_COHERENT_HEAP -#cmakedefine USE_CACHED_HEAP #cmakedefine USE_MANAGED_HEAP #cmakedefine USE_HOST_HEAP #cmakedefine USE_HIP_HOST_HEAP #cmakedefine USE_FUNC_CALL #cmakedefine USE_SINGLE_NODE -#cmakedefine USE_HOST_SIDE_HDP_FLUSH \ No newline at end of file +#cmakedefine USE_HOST_SIDE_HDP_FLUSH diff --git a/projects/rocshmem/scripts/build_configs/rc_single b/projects/rocshmem/scripts/build_configs/rc_single index 31eb7283c6..aceca7dbed 100755 --- a/projects/rocshmem/scripts/build_configs/rc_single +++ b/projects/rocshmem/scripts/build_configs/rc_single @@ -20,7 +20,6 @@ cmake \ -DUSE_DC=OFF \ -DUSE_IPC=OFF \ -DUSE_COHERENT_HEAP=OFF \ - -DUSE_CACHED_HEAP=OFF \ -DUSE_THREADS=OFF \ -DUSE_WF_COAL=OFF \ $src_path diff --git a/projects/rocshmem/scripts/build_configs/rc_single_single_node b/projects/rocshmem/scripts/build_configs/rc_single_single_node index 8bcf9fcb0d..a5223c379f 100755 --- a/projects/rocshmem/scripts/build_configs/rc_single_single_node +++ b/projects/rocshmem/scripts/build_configs/rc_single_single_node @@ -20,12 +20,10 @@ cmake \ -DUSE_DC=OFF \ -DUSE_IPC=ON \ -DUSE_COHERENT_HEAP=OFF \ - -DUSE_CACHED_HEAP=OFF \ -DUSE_THREADS=OFF \ -DUSE_WF_COAL=OFF \ -DUSE_SINGLE_NODE=ON \ -DUSE_HOST_SIDE_HDP_FLUSH=ON\ - -DROCM_PATH="/opt/rocm-5.4.2/"\ $src_path cmake --build . --parallel 8 cmake --install . diff --git a/projects/rocshmem/scripts/build_configs/rc_single_single_node_debug b/projects/rocshmem/scripts/build_configs/rc_single_single_node_debug index b535f1b853..c4e2477da9 100755 --- a/projects/rocshmem/scripts/build_configs/rc_single_single_node_debug +++ b/projects/rocshmem/scripts/build_configs/rc_single_single_node_debug @@ -20,7 +20,6 @@ cmake \ -DUSE_DC=OFF \ -DUSE_IPC=ON \ -DUSE_COHERENT_HEAP=OFF \ - -DUSE_CACHED_HEAP=OFF \ -DUSE_THREADS=OFF \ -DUSE_WF_COAL=OFF \ -DUSE_SINGLE_NODE=ON \ diff --git a/projects/rocshmem/src/gpu_ib/connection.cpp b/projects/rocshmem/src/gpu_ib/connection.cpp index 631f72f324..b6b2ae4dcd 100644 --- a/projects/rocshmem/src/gpu_ib/connection.cpp +++ b/projects/rocshmem/src/gpu_ib/connection.cpp @@ -256,7 +256,7 @@ void* Connection::buf_alloc([[maybe_unused]] struct ibv_pd* pd, if (use_gpu_mem) { void* dev_ptr; if (coherent_cq == 1) { -#if defined USE_COHERENT_HEAP || defined USE_CACHED_HEAP +#if defined USE_COHERENT_HEAP CHECK_HIP(hipMalloc(reinterpret_cast(&dev_ptr), size)); #else #ifdef HIP_SUPPORTS_MALLOC_UNCACHED diff --git a/projects/rocshmem/src/memory/heap_type.hpp b/projects/rocshmem/src/memory/heap_type.hpp index 0008317498..3f48edd11e 100644 --- a/projects/rocshmem/src/memory/heap_type.hpp +++ b/projects/rocshmem/src/memory/heap_type.hpp @@ -40,7 +40,7 @@ namespace rocshmem { #if defined USE_MANAGED_HEAP using HEAP_T = HeapMemory; -#elif defined USE_COHERENT_HEAP || defined USE_CACHED_HEAP +#elif defined USE_COHERENT_HEAP using HEAP_T = HeapMemory; #elif defined USE_HOST_HEAP using HEAP_T = HeapMemory; diff --git a/projects/rocshmem/src/util.hpp b/projects/rocshmem/src/util.hpp index 0d4dad806a..1d5cd9f9a6 100644 --- a/projects/rocshmem/src/util.hpp +++ b/projects/rocshmem/src/util.hpp @@ -126,7 +126,7 @@ extern const int gpu_clock_freq_mhz; __device__ __forceinline__ void __roc_inv() { asm volatile("buffer_wbinvl1;"); } __device__ __forceinline__ void __roc_flush() { -#if defined USE_COHERENT_HEAP || defined USE_CACHED_HEAP +#if defined USE_COHERENT_HEAP #if __gfx90a__ asm volatile("s_dcache_wb;"); asm volatile("buffer_wbl2;");