PyCuAmpcor: compile files as pure C++ when possible, fix Makefile

contrib/PyCuAmpcor/CMakeLists.txt

@@ -11,24 +11,26 @@ set(CMAKE_CUDA_STANDARD_REQUIRED TRUE)
 
 pybind11_add_module(PyCuAmpcor
     src/PyCuAmpcor.cpp
-    src/GDALImage.cu
+    src/GDALImage.cpp
     src/SConscript
-    src/cuAmpcorChunk.cu
-    src/cuAmpcorController.cu
-    src/cuAmpcorParameter.cu
-    src/cuArrays.cu
+    src/cuAmpcorChunk.cpp
+    src/cuAmpcorController.cpp
+    src/cuAmpcorParameter.cpp
+    src/cuArrays.cpp
     src/cuArraysCopy.cu
     src/cuArraysPadding.cu
     src/cuCorrFrequency.cu
     src/cuCorrNormalization.cu
     src/cuCorrNormalizationSAT.cu
-    src/cuCorrNormalizer.cu
+    src/cuCorrNormalizer.cpp
     src/cuCorrTimeDomain.cu
     src/cuDeramp.cu
     src/cuEstimateStats.cu
     src/cuOffset.cu
-    src/cuOverSampler.cu
+    src/cuOverSampler.cpp
     src/cuSincOverSampler.cu
+    src/cudaError.cpp
+    src/cudaUtil.cpp
     )
 target_include_directories(PyCuAmpcor PRIVATE
     src

contrib/PyCuAmpcor/README.md

@@ -59,8 +59,6 @@ You may also install PyCuAmpcor as a standalone package.
     # edit Makefile to provide the correct gdal include path and gpu architecture to NVCCFLAGS
     # call make to compile
     make
-    # install 
-    python3 setup.py install  
  ```
 
 ## 3. User Guide

contrib/PyCuAmpcor/src/GDALImage.cu → contrib/PyCuAmpcor/src/GDALImage.cpp

@@ -8,6 +8,7 @@
 #include "GDALImage.h"
 
 // dependencies
+#include <cuda_runtime.h>
 #include <iostream>
 #include "cudaError.h"
 

contrib/PyCuAmpcor/src/GDALImage.h

@@ -12,6 +12,7 @@
 #define __GDALIMAGE_H
 
 // dependencies
+#include <driver_types.h>
 #include <string>
 #include <gdal_priv.h>
 #include <cpl_conv.h>

contrib/PyCuAmpcor/src/Makefile

@@ -1,80 +1,47 @@
-PROJECT = CUAMPCOR
+CXX ?= g++
+NVCC ?= nvcc
 
-LDFLAGS =  -lcuda -lcudart -lcufft -lgdal
-CXXFLAGS = -std=c++11 -fpermissive -DNDEBUG -fPIC -shared
-NVCCFLAGS = -std=c++11 -m64 -DNDEBUG \
+CUDA_ROOT ?= $(dir $(shell which $(NVCC)))..
+
+LDFLAGS = -L$(CUDA_ROOT)/lib64 -L$(CUDA_ROOT)/lib64/stubs -lcuda -lcudart -lcufft -lgdal
+CXXFLAGS = -std=c++11 -fPIC -shared -I$(CUDA_ROOT)/include
+NVCCFLAGS = -std=c++11 -m64 \
 		-gencode arch=compute_35,code=sm_35 \
 		-gencode arch=compute_60,code=sm_60 \
 		-Xcompiler -fPIC -shared -Wno-deprecated-gpu-targets \
 		-ftz=false -prec-div=true -prec-sqrt=true \
 		-I/usr/include/gdal
 
-CXX=g++
-NVCC=nvcc
+CXXFLAGS += -O2 -DNDEBUG
+NVCCFLAGS += -O2 -DNDEBUG
+
+# pybind11 configuration
+PYTHON ?= python3
+PYTHON_CONFIG ?= python3-config
+PYTHON_EXT_SUFFIX := $(shell "$(PYTHON_CONFIG)" --extension-suffix)
+PYTHON_INCLUDES := $(shell "$(PYTHON)" -m pybind11 --includes) \
+                   $(shell "$(PYTHON_CONFIG)" --includes)
 
 DEPS = cudaUtil.h cudaError.h cuArrays.h GDALImage.h cuAmpcorParameter.h
 OBJS =  GDALImage.o cuArrays.o cuArraysCopy.o cuArraysPadding.o cuOverSampler.o \
+		cudaError.o cudaUtil.o \
 		cuSincOverSampler.o cuDeramp.o cuOffset.o \
 		cuCorrNormalization.o cuCorrNormalizationSAT.o cuCorrNormalizer.o \
 		cuAmpcorParameter.o cuCorrTimeDomain.o cuCorrFrequency.o \
 		cuAmpcorChunk.o cuAmpcorController.o cuEstimateStats.o
 
 all: pyampcor
 
-GDALImage.o: GDALImage.cu $(DEPS)
-	$(NVCC) $(NVCCFLAGS) -c -o $@ GDALImage.cu
-
-cuArrays.o: cuArrays.cu $(DEPS)
-	$(NVCC) $(NVCCFLAGS) -c -o $@ cuArrays.cu
-
-cuArraysCopy.o: cuArraysCopy.cu $(DEPS)
-	$(NVCC) $(NVCCFLAGS) -c -o $@ cuArraysCopy.cu
-
-cuArraysPadding.o: cuArraysPadding.cu $(DEPS)
-	$(NVCC) $(NVCCFLAGS) -c -o $@ cuArraysPadding.cu
-
-cuSincOverSampler.o: cuSincOverSampler.cu $(DEPS)
-	$(NVCC) $(NVCCFLAGS) -c -o $@ cuSincOverSampler.cu
-
-cuOverSampler.o: cuOverSampler.cu $(DEPS)
-	$(NVCC) $(NVCCFLAGS) -c -o $@ cuOverSampler.cu
-
-cuDeramp.o: cuDeramp.cu $(DEPS)
-	$(NVCC) $(NVCCFLAGS) -c -o $@ cuDeramp.cu
-
-cuOffset.o: cuOffset.cu $(DEPS)
-	$(NVCC) $(NVCCFLAGS) -c -o $@ cuOffset.cu
-
-cuCorrNormalization.o: cuCorrNormalization.cu $(DEPS)
-	$(NVCC) $(NVCCFLAGS) -c -o $@ cuCorrNormalization.cu
-
-cuCorrNormalizationSAT.o: cuCorrNormalizationSAT.cu $(DEPS)
-	$(NVCC) $(NVCCFLAGS) -c -o $@ cuCorrNormalizationSAT.cu
-
-cuCorrNormalizer.o: cuCorrNormalizer.cu $(DEPS)
-	$(NVCC) $(NVCCFLAGS) -c -o $@ cuCorrNormalizer.cu
-
-cuAmpcorParameter.o: cuAmpcorParameter.cu
-	$(NVCC) $(NVCCFLAGS) -c -o $@ cuAmpcorParameter.cu
-
-cuCorrTimeDomain.o: cuCorrTimeDomain.cu $(DEPS)
-	$(NVCC) $(NVCCFLAGS) -c -o $@ cuCorrTimeDomain.cu
-
-cuCorrFrequency.o: cuCorrFrequency.cu $(DEPS) cuCorrFrequency.h
-	$(NVCC) $(NVCCFLAGS) -c -o $@ cuCorrFrequency.cu
-
-cuAmpcorChunk.o: cuAmpcorChunk.cu cuAmpcorUtil.h $(DEPS)
-	$(NVCC) $(NVCCFLAGS) -c -o $@ cuAmpcorChunk.cu
-
-cuAmpcorController.o: cuAmpcorController.cu
-	$(NVCC) $(NVCCFLAGS) -c -o $@ cuAmpcorController.cu
+pyampcor: PyCuAmpcor$(PYTHON_EXT_SUFFIX)
 
-cuEstimateStats.o: cuEstimateStats.cu
-	$(NVCC) $(NVCCFLAGS) -c -o $@ cuEstimateStats.cu
+PyCuAmpcor$(PYTHON_EXT_SUFFIX): PyCuAmpcor.cpp $(OBJS)
+	$(CXX) $(CXXFLAGS) $(LDFLAGS) $(PYTHON_INCLUDES) $^ -o $@
 
+%.o: %.cu $(DEPS)
+	$(NVCC) $(NVCCFLAGS) -c -o $@ $<
 
-pyampcor: $(OBJS)
-	rm -f PyCuAmpcor.cpp && python3 setup.py build_ext --inplace
+%.o: %.cpp $(DEPS)
+	$(CXX) $(CXXFLAGS) -c -o $@ $<
 
 clean:
-	rm -rf *.o *so build *~  PyCuAmpcor.cpp *.dat
+	rm -rf *.o *.so build *~

contrib/PyCuAmpcor/src/SConscript

@@ -7,14 +7,15 @@ package = envPyCuAmpcor['PACKAGE']
 project = envPyCuAmpcor['PROJECT']
 build = envPyCuAmpcor['PRJ_LIB_DIR']
 install = envPyCuAmpcor['PRJ_SCONS_INSTALL'] + '/' + package + '/' + project
-listFiles = ['GDALImage.cu', 'cuArrays.cu', 'cuArraysCopy.cu',
+listFiles = ['GDALImage.cpp', 'cuArrays.cpp', 'cuArraysCopy.cu',
+             'cudaError.cpp', 'cudaUtil.cpp',
              'cuArraysPadding.cu', 'cuOverSampler.cu',
-             'cuSincOverSampler.cu', 'cuDeramp.cu',
+             'cuSincOverSampler.cpp', 'cuDeramp.cu',
              'cuOffset.cu', 'cuCorrNormalization.cu',
-             'cuCorrNormalizationSAT.cu', 'cuCorrNormalizer.cu',
-             'cuAmpcorParameter.cu', 'cuCorrTimeDomain.cu',
-             'cuAmpcorController.cu', 'cuCorrFrequency.cu',
-             'cuAmpcorChunk.cu', 'cuEstimateStats.cu']
+             'cuCorrNormalizationSAT.cu', 'cuCorrNormalizer.cpp',
+             'cuAmpcorParameter.cpp', 'cuCorrTimeDomain.cu',
+             'cuAmpcorController.cpp', 'cuCorrFrequency.cu',
+             'cuAmpcorChunk.cpp', 'cuEstimateStats.cu']
 
 lib = envPyCuAmpcor.SharedLibrary(target = 'PyCuAmpcor', source= listFiles, SHLIBPREFIX='')
 

contrib/PyCuAmpcor/src/cuAmpcorChunk.cu → contrib/PyCuAmpcor/src/cuAmpcorChunk.cpp

@@ -1,5 +1,8 @@
 #include "cuAmpcorChunk.h"
+
 #include "cuAmpcorUtil.h"
+#include <cufft.h>
+#include <iostream>
 
 /**
  * Run ampcor process for a batch of images (a chunk)

contrib/PyCuAmpcor/src/cuAmpcorController.cu → contrib/PyCuAmpcor/src/cuAmpcorController.cpp

@@ -12,6 +12,7 @@
 #include "cudaUtil.h"
 #include "cuAmpcorChunk.h"
 #include "cuAmpcorUtil.h"
+#include <cuda_runtime.h>
 #include <iostream>
 
 // constructor

contrib/PyCuAmpcor/src/cuArrays.cu → contrib/PyCuAmpcor/src/cuArrays.cpp

@@ -7,6 +7,9 @@
 // dependencies
 #include "cuArrays.h"
 #include "cudaError.h"
+#include <cuda_runtime.h>
+#include <fstream>
+#include <iostream>
 
 // allocate arrays in device memory
 template <typename T>

contrib/PyCuAmpcor/src/cuArrays.h

@@ -12,14 +12,9 @@
 #define __CUARRAYS_H
 
 // cuda dependencies
-#include <cuda.h>
 #include <driver_types.h>
 
-#include <iostream>
-#include <fstream>
-#include <cstdlib>
-#include <ctime>
-
+#include <string>
 
 template <typename T>
 class cuArrays{

contrib/PyCuAmpcor/src/cuCorrFrequency.cu

@@ -109,4 +109,4 @@ void cuArraysElementMultiplyConjugate(cuArrays<float2> *image1, cuArrays<float2>
     cudaKernel_elementMulConjugate<<<blockspergrid, threadsperblock, 0, stream>>>(image1->devData, image2->devData, size, coef );
     getLastCudaError("cuArraysElementMultiply error\n");
 } 
-//end of file
+//end of file

contrib/PyCuAmpcor/src/cuCorrFrequency.h

@@ -8,8 +8,8 @@
 #define __CUCORRFREQUENCY_H
 
 // dependencies
-#include "cudaUtil.h"
 #include "cuArrays.h"
+#include <cufft.h>
 
 class cuFreqCorrelator
 {
@@ -34,4 +34,4 @@ class cuFreqCorrelator
 };
 
 #endif //__CUCORRFREQUENCY_H
-// end of file
+// end of file

contrib/PyCuAmpcor/src/cuCorrNormalizer.h

@@ -13,7 +13,6 @@
 #define __CUNORMALIZER_H
 
 #include "cuArrays.h"
-#include "cudaUtil.h"
 
 /**
  * Abstract class interface for correlation surface normalization processor

contrib/PyCuAmpcor/src/cuOverSampler.h

@@ -12,7 +12,7 @@
 #define __CUOVERSAMPLER_H
 
 #include "cuArrays.h"
-#include "cudaUtil.h"
+#include <cufft.h>
 
 // FFT Oversampler for complex images
 class cuOverSamplerC2C

contrib/PyCuAmpcor/src/cudaError.cpp

@@ -0,0 +1,44 @@
+#include "cudaError.h"
+
+#include <cuda_runtime.h>
+#include <cufft.h>
+#include <stdio.h>
+#include <stdlib.h>
+
+#ifdef __DRIVER_TYPES_H__
+#ifndef DEVICE_RESET
+#define DEVICE_RESET cudaDeviceReset();
+#endif
+#else
+#ifndef DEVICE_RESET
+#define DEVICE_RESET
+#endif
+#endif
+
+template<typename T >
+void check(T result, char const *const func, const char *const file, int const line)
+{
+    if (result) {
+        fprintf(stderr, "CUDA error at %s:%d code=%d(%s) \n",
+                file, line, static_cast<unsigned int>(result), func);
+        DEVICE_RESET
+        // Make sure we call CUDA Device Reset before exiting
+        exit(EXIT_FAILURE);
+    }
+}
+
+template void check(cudaError_t, char const *const, const char *const, int const);
+template void check(cufftResult_t, char const *const, const char *const, int const);
+
+void __getLastCudaError(const char *errorMessage, const char *file, const int line)
+{
+    cudaError_t err = cudaGetLastError();
+
+    if (cudaSuccess != err)
+    {
+        fprintf(stderr, "%s(%i) : CUDA error : %s : (%d) %s.\n",
+                file, line, errorMessage, (int)err, cudaGetErrorString(err));
+        DEVICE_RESET
+        exit(EXIT_FAILURE);
+    }
+}

contrib/PyCuAmpcor/src/cudaError.h

@@ -10,51 +10,13 @@
 
 #pragma once
 
-#include <cstdlib>
-#include <cstdio>
-#include <cstring>
-#include <cufft.h>
 #include "debug.h"
-#include <cuda.h>
-
-
-#ifdef __DRIVER_TYPES_H__
-#ifndef DEVICE_RESET
-#define DEVICE_RESET cudaDeviceReset();
-#endif
-#else
-#ifndef DEVICE_RESET
-#define DEVICE_RESET
-#endif
-#endif
 
 template<typename T >
-void check(T result, char const *const func, const char *const file, int const line)
-{
-    if (result)
-    {
-
-        fprintf(stderr, "CUDA error at %s:%d code=%d(%s) \n",
-                file, line, static_cast<unsigned int>(result), func);
-        DEVICE_RESET
-        // Make sure we call CUDA Device Reset before exiting
-        exit(EXIT_FAILURE);
-    }
-}
+void check(T result, char const *const func, const char *const file, int const line);
 
 // This will output the proper error string when calling cudaGetLastError
-inline void __getLastCudaError(const char *errorMessage, const char *file, const int line)
-{
-    cudaError_t err = cudaGetLastError();
-
-    if (cudaSuccess != err)
-    {
-        fprintf(stderr, "%s(%i) : CUDA error : %s : (%d) %s.\n",
-                file, line, errorMessage, (int)err, cudaGetErrorString(err));
-        DEVICE_RESET
-        exit(EXIT_FAILURE);
-    }
-}
+void __getLastCudaError(const char *errorMessage, const char *file, const int line);
 
 // This will output the proper CUDA error strings in the event that a CUDA host call returns an error
 #ifdef CUDA_ERROR_CHECK