diff --git a/GPUSimulators/gpu/CMakeLists.txt b/GPUSimulators/gpu/CMakeLists.txt index bb35711..fa3d19b 100644 --- a/GPUSimulators/gpu/CMakeLists.txt +++ b/GPUSimulators/gpu/CMakeLists.txt @@ -1,34 +1,10 @@ -cmake_minimum_required(VERSION 4.0) -project(FiniteVolumeGPU LANGUAGES CXX CUDA) - -if(NOT DEFINED CMAKE_CUDA_ARCHITECTURES) - set(CMAKE_CUDA_ARCHITECTURES 52) -endif() +cmake_minimum_required(VERSION 3.21) +cmake_policy(VERSION 3.21.3...3.31.6) +project(FiniteVolumeGPU LANGUAGES CXX HIP CUDA) set(CMAKE_CXX_STANDARD 23) -include_directories(common) +set(CMAKE_EXPORT_COMPILE_COMMANDS ON) -add_library(cuda_comp SHARED - common/common.h - common/EulerCommon.h - common/limiters.h - common/SWECommon.h - cuda/EE2D_KP07_dimsplit.cu - cuda/SWE2D_FORCE.cu - cuda/SWE2D_HLL.cu - cuda/SWE2D_HLL2.cu - cuda/SWE2D_KP07.cu - cuda/SWE2D_KP07_dimsplit.cu - cuda/SWE2D_LxF.cu - cuda/SWE2D_WAF.cu - cuda/EE2D_KP07_dimsplit.cu - cuda/EE2D_KP07_dimsplit.cu - cuda/EE2D_KP07_dimsplit.cu) - -target_compile_definitions(cuda_comp PUBLIC - BLOCK_WIDTH - BLOCK_HEIGHT -) - -target_compile_features(cuda_comp PRIVATE cuda_std_14) +add_subdirectory(hip) +add_subdirectory(cuda) diff --git a/GPUSimulators/gpu/cuda/CMakeLists.txt b/GPUSimulators/gpu/cuda/CMakeLists.txt new file mode 100644 index 0000000..3acec87 --- /dev/null +++ b/GPUSimulators/gpu/cuda/CMakeLists.txt @@ -0,0 +1,16 @@ +if(NOT DEFINED CMAKE_CUDA_ARCHITECTURES) + set(CMAKE_CUDA_ARCHITECTURES 52) +endif() + +include_directories(${CMAKE_CURRENT_SOURCE_DIR}/include) + +file(GLOB MODELS "${CMAKE_CURRENT_SOURCE_DIR}/*.cu") + +add_library(cuda_comp MODULE ${MODELS}) + +target_compile_definitions(cuda_comp PUBLIC + BLOCK_WIDTH + BLOCK_HEIGHT +) + +target_compile_features(cuda_comp PRIVATE cuda_std_14) diff --git a/GPUSimulators/gpu/common/EulerCommon.h b/GPUSimulators/gpu/cuda/include/EulerCommon.h similarity index 100% rename from GPUSimulators/gpu/common/EulerCommon.h rename to GPUSimulators/gpu/cuda/include/EulerCommon.h diff --git a/GPUSimulators/gpu/common/SWECommon.h b/GPUSimulators/gpu/cuda/include/SWECommon.h similarity index 99% rename from GPUSimulators/gpu/common/SWECommon.h rename to GPUSimulators/gpu/cuda/include/SWECommon.h index b9218a6..bdaf156 100644 --- a/GPUSimulators/gpu/common/SWECommon.h +++ b/GPUSimulators/gpu/cuda/include/SWECommon.h @@ -1,5 +1,5 @@ /* -These CUDA functions implement different types of numerical flux +rThese CUDA functions implement different types of numerical flux functions for the shallow water equations Copyright (C) 2016, 2017, 2018 SINTEF Digital diff --git a/GPUSimulators/gpu/common/common.h b/GPUSimulators/gpu/cuda/include/common.h similarity index 99% rename from GPUSimulators/gpu/common/common.h rename to GPUSimulators/gpu/cuda/include/common.h index 9908dea..f3f9051 100644 --- a/GPUSimulators/gpu/common/common.h +++ b/GPUSimulators/gpu/cuda/include/common.h @@ -463,7 +463,7 @@ __device__ float reduce_max(float *data, unsigned int n) { // Reduce to "threads" elements sdata[tid] = FLT_MIN; for (unsigned int i = tid; i < n; i += threads) { - sdata[tid] = max(sdata[tid], dt_ctx.L[i]); + sdata[tid] = max(sdata[tid], data[i]); } __syncthreads(); diff --git a/GPUSimulators/gpu/common/limiters.h b/GPUSimulators/gpu/cuda/include/limiters.h similarity index 100% rename from GPUSimulators/gpu/common/limiters.h rename to GPUSimulators/gpu/cuda/include/limiters.h diff --git a/GPUSimulators/gpu/hip/CMakeLists.txt b/GPUSimulators/gpu/hip/CMakeLists.txt new file mode 100644 index 0000000..c87125d --- /dev/null +++ b/GPUSimulators/gpu/hip/CMakeLists.txt @@ -0,0 +1,18 @@ +include_directories(${CMAKE_CURRENT_SOURCE_DIR}/include) + +execute_process(COMMAND hipconfig --rocmpath + OUTPUT_VARIABLE rocm_path) +message(STATUS "ROCm SDK path: ${rocm_path}") +set(CMAKE_HIP_FLAGS "${CMAKE_HIP_FLAGS} -isystem ${rocm_path}/include") + + +file(GLOB MODELS "${CMAKE_CURRENT_SOURCE_DIR}/*.hip") + +add_library(hip_comp ${MODELS}) + +target_compile_definitions(hip_comp PUBLIC + BLOCK_WIDTH + BLOCK_HEIGHT +) + +target_compile_features(hip_comp PRIVATE cxx_std_23)