Migrate max pooling

include/cudanet.cuh

@@ -18,7 +18,7 @@
 #include "dense.hpp"
 #include "input.cuh"
 #include "layer.hpp"
-#include "max_pooling.cuh"
+#include "max_pooling.hpp"
 #include "output.cuh"
 
 // Models

include/layers/max_pooling.hpp

@@ -18,7 +18,7 @@ class MaxPooling2d : public SequentialLayer, public TwoDLayer {
     );
     ~MaxPooling2d();
 
-    float* forward(const float* d_input);
+    float* forward(const float* input);
 
     /**
      * @brief Get output size
@@ -45,9 +45,17 @@ class MaxPooling2d : public SequentialLayer, public TwoDLayer {
 
     shape2d outputSize;
 
-    float* d_output;
-
     Activation* activation;
+
+    float* forwardCPU(const float* input);
+
+#ifdef USE_CUDA
+    float* d_output;
+    float* forwardCUDA(const float* d_input);
+
+    void initCUDA();
+    void delCUDA();
+#endif
 };
 
 }  // namespace CUDANet::Layers

src/backends/cuda/layers/max_pooling.cu

@@ -0,0 +1,38 @@
+#include "cuda_helper.cuh"
+#include "max_pooling.hpp"
+#include "pooling.cuh"
+
+using namespace CUDANet::Layers;
+
+void MaxPooling2d::initCUDA() {
+        d_output = nullptr;
+    CUDA_CHECK(cudaMalloc(
+        (void**)&d_output,
+        sizeof(float) * outputSize.first * outputSize.second * nChannels
+    ));
+}
+
+void MaxPooling2d::delCUDA() {
+    cudaFree(d_output);
+}
+
+
+float* MaxPooling2d::forwardCUDA(const float* d_input) {
+    dim3 block(8, 8, 8);
+    dim3 grid(
+        (outputSize.first + block.x - 1) / block.x,
+        (outputSize.second + block.y - 1) / block.y,
+        (nChannels + block.z - 1) / block.z
+    );
+
+    Kernels::max_pooling<<<grid, block>>>(
+        d_input, d_output, inputSize, outputSize, nChannels, poolingSize,
+        stride, padding
+    );
+    CUDA_CHECK(cudaGetLastError());
+
+    activation->activate(d_output);
+    CUDA_CHECK(cudaDeviceSynchronize());
+
+    return d_output;
+}

src/layers/max_pooling.cpp

@@ -1,6 +1,5 @@
-#include "cuda_helper.cuh"
-#include "max_pooling.cuh"
-#include "pooling.cuh"
+#include "max_pooling.hpp"
+#include <stdexcept>
 
 using namespace CUDANet::Layers;
 
@@ -30,38 +29,31 @@ MaxPooling2d::MaxPooling2d(
         activationType, outputSize.first * outputSize.second * nChannels
     );
 
-    d_output = nullptr;
-    CUDA_CHECK(cudaMalloc(
-        (void**)&d_output,
-        sizeof(float) * outputSize.first * outputSize.second * nChannels
-    ));
+    #ifdef USE_CUDA
+    initCUDA();
+#endif
 }
 
 MaxPooling2d::~MaxPooling2d() {
-    cudaFree(d_output);
+#ifdef USE_CUDA
+    delCUDA();
+#endif
     delete activation;
 }
 
-float* MaxPooling2d::forward(const float* d_input) {
-    dim3 block(8, 8, 8);
-    dim3 grid(
-        (outputSize.first + block.x - 1) / block.x,
-        (outputSize.second + block.y - 1) / block.y,
-        (nChannels + block.z - 1) / block.z
-    );
-
-    Kernels::max_pooling<<<grid, block>>>(
-        d_input, d_output, inputSize, outputSize, nChannels, poolingSize,
-        stride, padding
-    );
-    CUDA_CHECK(cudaGetLastError());
-
-    activation->activate(d_output);
-    CUDA_CHECK(cudaDeviceSynchronize());
-
-    return d_output;
+float* MaxPooling2d::forwardCPU(const float* input) {
+    throw std::logic_error("Not implemented");
 }
 
+float* MaxPooling2d::forward(const float* input) {
+#ifdef USE_CUDA
+    return forwardCUDA(input);
+#else
+    return forwardCPU(input);
+#endif
+}
+
+
 int MaxPooling2d::getOutputSize() {
     return outputSize.first * outputSize.second * nChannels;
 }

test/cuda/layers/test_max_pooling.cu

@@ -3,7 +3,7 @@
 
 #include <vector>
 
-#include "max_pooling.cuh"
+#include "max_pooling.hpp"
 
 class MaxPoolingLayerTest : public ::testing::Test {
   protected:

test/model/test_model.cpp

@@ -2,7 +2,7 @@
 
 #include "conv2d.cuh"
 #include "dense.hpp"
-#include "max_pooling.cuh"
+#include "max_pooling.hpp"
 #include "model.hpp"
 
 class ModelTest : public ::testing::Test {