Various bug fixes for older compilers (#588)

* Various bug fixes for older compilers

* Stop CMake if g++ is too old

Author: cliffburdick

CMakeLists.txt

@@ -78,6 +78,13 @@ rapids_cmake_write_version_file(include/version_config.h)
 set(CMAKE_CXX_STANDARD 17)
 set(CUDA_CXX_STANDARD 17)
 
+if ("${CMAKE_CXX_COMPILER_ID}" MATCHES "GNU")
+    execute_process(COMMAND ${CMAKE_CXX_COMPILER} -dumpversion OUTPUT_VARIABLE GCC_VERSION)
+    if (NOT GCC_VERSION VERSION_GREATER 8)
+        message(FATAL_ERROR "${PROJECT_NAME} requires g++ 9 or higher if using g++ host compiler")
+    endif()
+endif()
+
 # CPM is required for all package management
 include(public/cpm-cmake/cmake/CPM.cmake)
 # Helper for selecting build type

include/matx/core/allocator.h

@@ -102,7 +102,7 @@ struct MemTracker {
   }
 
   template <typename StreamType>
-  auto deallocate_internal(void *ptr, StreamType st) {
+  auto deallocate_internal(void *ptr, [[maybe_unused]] StreamType st) {
     MATX_NVTX_START("", matx::MATX_NVTX_LOG_INTERNAL)
 
     std::unique_lock lck(memory_mtx);

include/matx/core/defines.h

@@ -68,4 +68,7 @@ static_assert(false, "Must choose either 64-bit or 32-bit index mode");
     #define __MATX_INLINE__ inline
 #endif
 
+// std::ceil is not constexpr until C++23
+#define MATX_ROUND_UP(N, S) ((((N) + (S) - 1) / (S)) * (S))
+
 }

include/matx/kernels/conv.cuh

@@ -232,9 +232,7 @@ __global__ void Conv2D(OutType d_out, InType1 d_in1, InType2 d_in2,
 
   constexpr int Rank = OutType::Rank();
 
-  constexpr int type2off = 
-    std::ceil(static_cast<double>(sizeof(in1type) * SIGNAL_CHUNK_Y * SIGNAL_CHUNK_X) / sizeof(in2type) * 
-              sizeof(in2type));
+  constexpr int type2off = MATX_ROUND_UP(sizeof(in1type) * SIGNAL_CHUNK_Y * SIGNAL_CHUNK_X, sizeof(in2type));
   __shared__ char shared_buf[type2off + sizeof(in2type) * FILTER_SHARED_CHUNK_Y * FILTER_SHARED_CHUNK_X];
 
   // __shared__ Uninitialized<in1type> s_signal[SIGNAL_CHUNK_Y][SIGNAL_CHUNK_X];

include/matx/transforms/fft/fft_fftw.h

@@ -261,8 +261,8 @@ struct FftFFTWparams_t {
 
     auto exec_plans = [&](typename OutputTensor::value_type *out_ptr, 
                           typename InputTensor::value_type *in_ptr) {
-      fftwf_plan plan{};
       if constexpr (fp32) {
+        fftwf_plan plan{};
         if constexpr (DeduceFFTTransformType<OutputTensor, InputTensor>() == FFTType::C2C) {
           plan  = fftwf_plan_many_dft( params.fft_rank, 
                                       params.n, 
@@ -312,6 +312,7 @@ struct FftFFTWparams_t {
         fftwf_cleanup();         
       }
       else {
+        fftw_plan plan{};
         if constexpr (DeduceFFTTransformType<OutputTensor, InputTensor>() == FFTType::Z2Z) {
           plan  = fftw_plan_many_dft( params.fft_rank, 
                                       params.n,