The tracking system now working also on the GPU

CMakeLists.txt

@@ -11,7 +11,7 @@ include_directories(include
   ${CUDA_INCLUDE_DIRS})
 
 set(CUDA_NVCC_FLAGS "${CUDA_NVCC_FLAGS} -std=c++11 -arch=sm_30" )
-cuda_add_library(cu_ball_track
+cuda_add_library(cu_ball_track SHARED
   src/cuda/img_proc.cu
   )
 
@@ -19,6 +19,7 @@ add_library(ball_tracking SHARED
   src/img_proc.cpp
   src/utils.cpp
   src/tracker.cpp
+  src/cuda/tracker.cpp
   )
 target_link_libraries(ball_tracking
   ${OpenCV_LIBS}

examples/CMakeLists.txt

@@ -34,3 +34,11 @@ target_link_libraries(track
   ball_tracking
   ${Boost_LIBRARIES}
   )
+
+add_executable(gpu_track
+  tracking/gpu_track.cpp
+  )
+target_link_libraries(gpu_track
+  ball_tracking
+  ${Boost_LIBRARIES}
+  )

examples/img_proc/log_reg_approach.cpp

@@ -3,11 +3,22 @@
 #include <ball_tracking/utils.hpp>
 #include <opencv2/opencv.hpp>
 #include <opencv2/highgui/highgui.hpp>
+#include <opencv2/core.hpp>
 #include <iostream>
 #include <boost/program_options.hpp>
 #include <json.hpp>
 #include <chrono>
 
+#define USE_GPU
+
+#ifdef USE_GPU
+#include <opencv2/cudaarithm.hpp>
+#include <opencv2/cudaimgproc.hpp>
+#include <ball_tracking/cuda/img_proc.hpp>
+
+using namespace cv::cuda;
+#endif
+
 using namespace cv;
 using namespace std;
 using namespace ball_tracking;
@@ -47,27 +58,44 @@ int main(int argc, char** argv) {
     string fname = vm["input"].as<string>();
     Mat img = imread(fname, CV_LOAD_IMAGE_COLOR);
 
-    imshow("Color", back);
-    waitKey(0); 
+    /*imshow("Color", back);
+    waitKey(0);*/ 
     //Calculate the data with coefs
+#ifdef USE_GPU
+    cout << "Using the GPU for computation" << endl;
+    //1) Load images and weights to GPU
+    GpuMat gpu_img(img), gpu_bck(back), gpu_w(model_coef);
+    GpuMat gpu_result;
+    //2) Do the computation
+    auto start_t = std::chrono::steady_clock::now();
+    ball_tracking::cuda::quadf_log_reg(gpu_img, gpu_bck, gpu_w, gpu_result);
+    auto end_t = std::chrono::steady_clock::now();
+    //3) Load back to CPU
+    Mat im_proc; 
+    gpu_result.download(im_proc);
+#else
+    cout << "Using the CPU for computation" << endl;
     auto start_t = std::chrono::steady_clock::now();
     Mat im_proc = quadf_log_reg(img, back, model_coef);
     auto end_t = std::chrono::steady_clock::now();
+#endif
     std::cout << "Computation took "
               << std::chrono::duration_cast<std::chrono::milliseconds>(end_t - start_t).count() << 
               " milliseconds" << endl;
+    //4) Turn it into a probability image for easier interpretation
+    for (auto it=im_proc.begin<double>(); it != im_proc.end<double>(); ++it) {
+      *it = 1.0 / (1.0 + exp(-*it));
+    }
+    imshow("Likelihood", im_proc);
+    waitKey(0);
+
     /*cout << "pix data:" << im_proc.at<double>(340, 290) << endl 
       << im_proc.at<double>(342, 291) << endl 
       << im_proc.at<double>(290, 340) << endl
       << im_proc.at<double>(2, 2)   << endl;*/
-    MatIterator_<double> it_ans = im_proc.begin<double>(),end=im_proc.end<double>();
-    //MatIterator_<Vec3b> it=im_proc.begin<Vec3b>(), end=im_proc.end<Vec3b>();
-    while(it_ans != end){
-      //cout << "pix data:" << *it_ans << endl;	
-      double prob = 1.0 / (1.0 + exp(-*it_ans));
-     //*it_ans = 255*prob;
-      *it_ans = (prob > 0.95) ? 255 : 0;
-      it_ans++;
+    //5) Binarize
+    for (auto it=im_proc.begin<double>(); it != im_proc.end<double>(); ++it) {
+      *it = (*it > 0.95) ? 255 : 0;
     }
     imshow("Color", im_proc);
     waitKey(0); 

examples/tracking/gpu_track.cpp

+
+#include <ball_tracking/tracker.hpp>
+#include <ball_tracking/cuda/tracker.hpp>
+#include <ball_tracking/utils.hpp>
+#include <opencv2/opencv.hpp>
+#include <opencv2/core.hpp>
+#include <opencv2/highgui/highgui.hpp>
+#include <iostream>
+#include <boost/program_options.hpp>
+#include <json.hpp>
+#include <chrono>
+
+using namespace cv;
+using namespace std;
+using namespace ball_tracking;
+using json = nlohmann::json;
+
+int main(int argc, char** argv) {
+  try {
+    namespace po = boost::program_options;
+    po::options_description desc("Options");
+    desc.add_options()
+      ("help", "Produce help message")
+      ("input,i", po::value<string>(), "path to the input video")
+      ("output,o", po::value<string>(), "path to the output video")
+      ("conf,c", po::value<string>(), "path to the JSON configuration");
+    po::variables_map vm;
+    po::store(po::parse_command_line(argc, argv, desc), vm);
+    po::notify(vm);
+
+    if (vm.count("help")) {
+      cout << desc << endl;
+      return 0;
+    }
+    if (!vm.count("input")) {
+      cerr << "Error: You should provide the video to process" << endl;
+      return 1;
+    }
+
+    json jobj = load_json(vm["conf"].as<string>());
+    ball_tracking::cuda::BallLogLikelihood bll(jobj.at("ball_log_lh"));
+    Binarizer bin(jobj.at("binarizer"));
+    ball_tracking::cuda::Binarizer gpu_bin(jobj.at("binarizer"));
+    FindBallBlob bf(jobj.at("blob_detection"));
+
+    string fname = vm["input"].as<string>();
+    VideoCapture in_video(fname);
+    Size in_size((int)in_video.get(CV_CAP_PROP_FRAME_WIDTH), 
+        (int)in_video.get(CV_CAP_PROP_FRAME_HEIGHT));
+    int ex = static_cast<int>(in_video.get(CV_CAP_PROP_FOURCC)); // Use same input codec
+    // Transform from int to char via Bitwise operators
+    char codec[] = {(char)(ex & 0XFF) , 
+      (char)((ex & 0XFF00) >> 8),
+      (char)((ex & 0XFF0000) >> 16),
+      (char)((ex & 0XFF000000) >> 24), 0};
+    cout << "Input video codec: " << codec << " fps: " << in_video.get(CV_CAP_PROP_FPS) << endl;
+
+    string outname = vm["output"].as<string>();
+    Size out_size(2*in_size.width, 2*in_size.height);
+    VideoWriter out_video(outname, ex,//VideoWriter::fourcc(codec[0], codec[1], codec[2], codec[3]),
+        in_video.get(CV_CAP_PROP_FPS), out_size);
+
+    if (!in_video.isOpened()) {
+      cerr << "Input video could not be opened: " << fname << endl;
+      return -1;
+    }
+    if(!out_video.isOpened()) {
+      cerr << "Output video could not be opened: " << outname << endl;
+      return -1;
+    }
+
+    Mat img, l_lh, bin_img, tmp;
+    Mat res(out_size, CV_8UC3);
+    vector<double> llh_time, bin_time, blob_time, proc_time, in_time, draw_time;
+    //cv::cuda::Stream stream;
+    while (true) {
+      //1) Read the image
+      auto start_t = std::chrono::steady_clock::now();
+      auto st_0 = start_t;
+      in_video >> img;
+      if (img.empty()) break;
+      cv::cuda::GpuMat gpu_img, gpu_l_lh, gpu_bin_img;
+      gpu_img.upload(img);
+      auto end_t = std::chrono::steady_clock::now();
+      in_time.push_back(std::chrono::duration_cast<std::chrono::microseconds>(end_t - start_t).count()); 
+
+      //2) Compute log likelihood image (In GPU)
+      start_t = std::chrono::steady_clock::now();
+      bll(gpu_img, gpu_l_lh);
+      end_t = std::chrono::steady_clock::now();
+      llh_time.push_back(std::chrono::duration_cast<std::chrono::microseconds>(end_t - start_t).count());
+
+      //2.1) Binarize (In GPU)
+      start_t = std::chrono::steady_clock::now();
+      gpu_bin(gpu_l_lh, gpu_bin_img);
+      //2.2) Download the image back to the CPU
+      gpu_bin_img.download(bin_img);
+      end_t = std::chrono::steady_clock::now();
+      bin_time.push_back(std::chrono::duration_cast<std::chrono::microseconds>(end_t - start_t).count());
+
+      //3) Run blob detection algorithm (In CPU)
+      start_t = std::chrono::steady_clock::now();
+      auto key_pts = bf(bin_img);
+      end_t = std::chrono::steady_clock::now();
+      blob_time.push_back(std::chrono::duration_cast<std::chrono::microseconds>(end_t - start_t).count());
+    
+      proc_time.push_back(llh_time.back() + bin_time.back() + blob_time.back());
+      cout << "N: " << key_pts.size() << ", FPT: " << proc_time.back() << " us, ";
+      cout << "llh: " << llh_time.back() << " us, bin: " << bin_time.back() << " us, blob: " << blob_time.back() << " us, ";
+
+      //4) Compute output image
+      start_t = std::chrono::steady_clock::now();
+      gpu_l_lh.download(l_lh);
+      drawKeypoints(img, key_pts, res(Rect(0,0,in_size.width,in_size.height)),
+          Scalar(255,255,0), DrawMatchesFlags::DRAW_RICH_KEYPOINTS);
+      normalize(l_lh, tmp, 0, 1, NORM_MINMAX);
+      tmp.convertTo(tmp, CV_8U, 255);
+      drawKeypoints(tmp, key_pts, res(Rect(in_size.width,0,in_size.width,in_size.height)), 
+          Scalar(255,255,0), DrawMatchesFlags::DRAW_RICH_KEYPOINTS);
+      //res.adjustROI(0,0,in_size.width,0);
+      drawKeypoints(bin_img, key_pts, res(Rect(0,in_size.height,in_size.width,in_size.height)), 
+          Scalar(255,255,0), DrawMatchesFlags::DRAW_RICH_KEYPOINTS);
+      end_t = std::chrono::steady_clock::now();
+      draw_time.push_back(std::chrono::duration_cast<std::chrono::milliseconds>(end_t - start_t).count());
+      cout << "Draw: " << draw_time.back() << " ms, ";
+    
+      //5) Write output
+      start_t = std::chrono::steady_clock::now();
+      out_video << res;
+
+      //6) Display output
+      imshow("Output", res);
+      int k = waitKey(5) & 0xff;
+      end_t = std::chrono::steady_clock::now();
+      cout << " TT: " << std::chrono::duration_cast<std::chrono::milliseconds>(end_t - st_0).count() 
+        << " ms" << endl;
+      if (k==27) break;
+    }
+    return 0;
+  } catch (std::exception& ex) {
+    cerr << "Exception: " << ex.what() << endl;
+    return 1;
+  }
+}

examples/tracking/track.cpp

@@ -68,7 +68,7 @@ int main(int argc, char** argv) {
 
     Mat img, l_lh, tmp;
     Mat res(out_size, CV_8UC3);
-    vector<double> llh_time, blob_time, proc_time, in_time, draw_time;
+    vector<double> llh_time, bin_time, blob_time, proc_time, in_time, draw_time;
     while (true) {
       //1) Read the image
       auto start_t = std::chrono::steady_clock::now();
@@ -76,23 +76,29 @@ int main(int argc, char** argv) {
       in_video >> img;
       if (img.empty()) break;
       auto end_t = std::chrono::steady_clock::now();
-      in_time.push_back(std::chrono::duration_cast<std::chrono::milliseconds>(end_t - start_t).count()); 
+      in_time.push_back(std::chrono::duration_cast<std::chrono::microseconds>(end_t - start_t).count()); 
 
       //2) Compute log likelihood image
       start_t = std::chrono::steady_clock::now();
       l_lh = bll(img);
       end_t = std::chrono::steady_clock::now();
-      llh_time.push_back(std::chrono::duration_cast<std::chrono::milliseconds>(end_t - start_t).count());
+      llh_time.push_back(std::chrono::duration_cast<std::chrono::microseconds>(end_t - start_t).count());
 
-      //3) Run blob detection algorithm
+      //2.1) Binarize
       start_t = std::chrono::steady_clock::now();
       auto bin_img = bin(l_lh);
+      end_t = std::chrono::steady_clock::now();
+      bin_time.push_back(std::chrono::duration_cast<std::chrono::microseconds>(end_t - start_t).count());
+
+      //3) Run blob detection algorithm
+      start_t = std::chrono::steady_clock::now();
       auto key_pts = bf(bin_img);
       end_t = std::chrono::steady_clock::now();
-      blob_time.push_back(std::chrono::duration_cast<std::chrono::milliseconds>(end_t - start_t).count());
+      blob_time.push_back(std::chrono::duration_cast<std::chrono::microseconds>(end_t - start_t).count());
     
-      proc_time.push_back(llh_time.back() + blob_time.back());
-      cout << "N: " << key_pts.size() << ", FPT: " << proc_time.back() << " ms, ";
+      proc_time.push_back(llh_time.back() + bin_time.back() + blob_time.back());
+      cout << "N: " << key_pts.size() << ", FPT: " << proc_time.back() << " us, ";
+      cout << "llh: " << llh_time.back() << " us, bin: " << bin_time.back() << " us, blob: " << blob_time.back() << " us, ";
 
       //4) Compute output image
       start_t = std::chrono::steady_clock::now();

include/ball_tracking/cuda/img_proc.hpp

+#ifndef BALL_TRACKING_CUDA_IMG_PROC
+#define BALL_TRACKING_CUDA_IMG_PROC
+
+#include <opencv2/opencv.hpp>
+#include <memory>
+#include <functional>
+
+namespace ball_tracking {
+
+  namespace cuda {
+
+    /**
+     * @brief Applies a pixel-wise logistic regression with quadratic features to the source
+     * image and returns a single-channel image with the log-probabilities in the GPU
+     *
+     * @param[in] src The source image on the GPU
+     * @param[in] bkg A background image (Without the ball) on the GPU
+     * @param[in] weights Vector of weights of logistic regression
+     * @param[out] dst The destination image in the GPU
+     * @param[in] stream The stream of execution in the GPU
+     */
+    void quadf_log_reg(const cv::cuda::GpuMat& src, const cv::cuda::GpuMat& bkg, 
+        const cv::cuda::GpuMat& weights, cv::cuda::GpuMat& dst, 
+        cv::cuda::Stream& stream = cv::cuda::Stream::Null());
+  };
+
+};
+
+#endif

include/ball_tracking/cuda/tracker.hpp

+#ifndef BALL_TRACKING_CUDA_TRACKER
+#define BALL_TRACKING_CUDA_TRACKER
+
+#include <opencv2/opencv.hpp>
+#include <memory>
+#include <functional>
+#include <json.hpp>
+
+namespace ball_tracking {
+
+  namespace cuda {
+
+    /**
+     * @brief Data type for an Image pre-processing function.
+     */
+    using preproc = std::function<void(const cv::cuda::GpuMat& src, cv::cuda::GpuMat& dst, cv::cuda::Stream& stream)>;
+
+    /**
+     * @brief Produces a single channel image highlighting the possible position of the ball
+     * according to a likelihood function
+     */
+    class BallLogLikelihood {
+      private:
+        class Impl;
+        std::unique_ptr<Impl> _impl;
+      public:
+        /**
+         * @brief Creates an object with given configuration parameters
+         */
+        BallLogLikelihood(const nlohmann::json& params);
+        ~BallLogLikelihood();
+
+        /**
+         * @brief Produces a single channel image representing the log likelihood of each
+         * pixel being part of the ball
+         *
+         * @params[in] src The source image
+         * @params[out] dst  A new single channel image with the same dimensions of the input image,
+         * where every pixel contains the log likelihood of the corresponding pixel in the
+         * input image corresponding to a table tennis ball
+         * @params[in] stream The stream of execution on the GPU
+         */
+        void operator()(const cv::cuda::GpuMat& src, cv::cuda::GpuMat& dst, cv::cuda::Stream& stream = cv::cuda::Stream::Null());
+    };
+
+    /**
+     * @brief Thresholds an image
+     */
+    class Binarizer {
+      private:
+        class Impl;
+        std::unique_ptr<Impl> _impl;
+      public:
+        Binarizer(const nlohmann::json& params);
+        ~Binarizer();
+
+        /**
+         * @brief Returns an 8 bit single channel binarized image from a given single channel
+         * real valued image. The image contains only 0 or 255 values in each pixel.
+         *
+         * @param[in] src A single channel image where threshold is to be applied
+         * @returns a single 8 bit channel image with values 0 or 255 only.
+         */
+        void operator()(const cv::cuda::GpuMat& src, cv::cuda::GpuMat& dst, cv::cuda::Stream& stream = cv::cuda::Stream::Null());
+    };
+
+    /**
+     * @brief Find candidate ball positions if any in a likelihood image 
+     */
+    class FindBallBlob {
+      private:
+        class Impl;
+        std::unique_ptr<Impl> _impl;
+      public:
+        FindBallBlob(const nlohmann::json& params);
+        ~FindBallBlob();
+
+        /**
+         * @brief Returns all candidate locations of a ball in a single channel
+         * likelihood image.
+         *
+         * @param[in] src A binary (thresholded) image segmenting the interesting parts
+         * @returns a vector of Key Points with the locations and other optional
+         * properties of the ball blobs found in the image
+         */
+        std::vector<cv::KeyPoint> operator()(cv::InputArray src);
+    };
+
+  };
+};
+
+#endif

include/ball_tracking/img_proc.hpp

@@ -34,22 +34,6 @@ namespace ball_tracking {
    */
   cv::Mat quadf_log_reg(cv::InputArray src, cv::InputArray bkg, cv::InputArray weights);
 
-  namespace cuda {
-
-    /**
-     * @brief Applies a pixel-wise logistic regression with quadratic features to the source
-     * image and returns a single-channel image with the log-probabilities in the GPU
-     *
-     * @param[in] src The source image on the GPU
-     * @param[in] bkg A background image (Without the ball) on the GPU
-     * @param[in] weights Vector of weights of logistic regression
-     * @param[out] dst The destination image in the GPU
-     * @param[in] stream The stream of execution in the GPU
-     */
-    void quadf_log_reg(const cv::cuda::GpuMat& src, const cv::cuda::GpuMat& bkg, 
-        const cv::cuda::GpuMat weights, cv::cuda::GpuMat& dst, cv::cuda::Stream& stream);
-  };
-
 };
 
 #endif

src/cuda/img_proc.cu

 
 #include "opencv2/cudev.hpp"
 #include "opencv2/opencv_modules.hpp"
-#include <ball_tracking/img_proc.hpp>
+#include <ball_tracking/cuda/img_proc.hpp>
 
 using namespace cv;
 using namespace cv::cuda;
@@ -14,24 +14,25 @@ namespace ball_tracking {
         PtrStepSz<double> dst) {
       const unsigned int x = blockIdx.x*blockDim.x + threadIdx.x;
       const unsigned int y = blockIdx.y*blockDim.y + threadIdx.y;
-      const unsigned int n = x + blockIdx.x*blockDim.x*y;
+      const unsigned int n = threadIdx.x + blockDim.x*threadIdx.y;
       
       __shared__ double w[28];
       if (n<28) w[n] = weights(0,n);
       __syncthreads();
 
-      if (x < src.rows && y < src.cols) {
+      if (y < src.rows && x < src.cols) {
         double sum = 0.0;
-        uchar3 it = src(x,y), it_bkg = bkg(x,y);
+        uchar3 it = src(y,x), it_bkg = bkg(y,x);
         const double lfeat[7] = {it.x/255.0, it.y/255.0, it.z/255.0, 
           it_bkg.x/255.0, it_bkg.y/255.0, it_bkg.z/255.0, 1.0};
-        const double* wptr = w;
+        unsigned int k=0;
         for (unsigned int i=0; i<7; i++) {
-          for (unsigned int j=i; j<7; j++, wptr++) {
-            sum += (*wptr) * lfeat[i] * lfeat[j];
+          for (unsigned int j=i; j<7; j++, k++) {
+            sum += w[k] * lfeat[i] * lfeat[j];
           }
         }
         dst(y,x) = sum;
+        //dst(y,x) = (lfeat[0] + lfeat[1] + lfeat[2]) / 3.0;
       }
     }
   };
@@ -42,7 +43,7 @@ namespace ball_tracking {
      * is a vector of 28 dimensions
      */
     void quadf_log_reg(const GpuMat& src, const GpuMat& bkg,
-        const GpuMat weights, GpuMat& dst, Stream& stream = Stream::Null()) {
+        const GpuMat& weights, GpuMat& dst, Stream& stream) {
       CV_Assert(src.rows==bkg.rows && src.cols==bkg.cols);
       dim3 block(32,8);
       dim3 grid((src.cols + block.x - 1)/block.x,

src/cuda/tracker.cpp

+
+#include <opencv2/opencv.hpp>
+#include <memory>
+#include <functional>
+#include <json.hpp>
+#include <ball_tracking/cuda/img_proc.hpp>
+#include <ball_tracking/cuda/tracker.hpp>
+#include <ball_tracking/utils.hpp>
+#include <opencv2/core.hpp>
+
+using namespace cv::cuda;
+using namespace cv;
+using json = nlohmann::json;
+using namespace std;
+
+namespace ball_tracking {
+
+  namespace cuda {
+
+    namespace {
+
+      /**
+       * Color and background based logistic regression
+       */
+      class CB_log_reg {
+        private:
+          GpuMat bkg; //!< Background image
+          GpuMat weights; //!< Model weights
+          vector<Ptr<Filter>> pre; //!< Filtering functions for the images
+        public:
+   
+          void pre_chain(const GpuMat& _src, GpuMat& _dst, Stream& stream = Stream::Null()) {
+            GpuMat tmp = _src;
+            for (auto p : pre) {
+              p->apply(tmp, tmp, stream);
+            }
+            _dst = tmp;
+          }
+   
+          void operator()(const GpuMat& _src, GpuMat& dst, Stream& stream = Stream::Null()) {
+            GpuMat src;
+            pre_chain(_src, src, stream);
+            if (bkg.empty()) {
+              bkg = src;
+            }
+            quadf_log_reg(src, bkg, weights, dst, stream);
+          }
+
+          CB_log_reg(const json& conf, Stream& stream = Stream::Null()) {
+            Mat w = json2cvmat(conf.at("weights"));
+            weights.upload(w, stream);
+            if (conf.count("gauss_smooth")) {
+              auto gs = conf.at("gauss_smooth");
+              unsigned int ksize = gs.at("size");
+              double sigma = gs.at("sigma");
+              pre.push_back(createGaussianFilter(CV_8UC3, CV_8UC3, Size(ksize,ksize), sigma, sigma));
+            }
+            if (conf.count("background")) {
+              Mat host_bkg = imread(conf.at("background"), CV_LOAD_IMAGE_COLOR);
+              if (!host_bkg.data) {
+                throw std::logic_error("File not found for background image");
+              }
+              bkg.upload(host_bkg, stream);
+              pre_chain(bkg, bkg, stream);
+            }
+          }
+      };
+    };
+
+    class BallLogLikelihood::Impl {
+      public:
+        preproc bll;
+
+        Impl(const json& config) {
+          const string& type = config.at("type");
+          const json& conf = config.at("conf");
+          if (type == "cb_log_reg") {
+            bll = CB_log_reg(conf);
+          } else {
+            throw std::logic_error("Type of ball log likelihood algorithm not recognized");
+          }
+        }
+    };
+
+    BallLogLikelihood::BallLogLikelihood(const nlohmann::json& params) {
+      _impl = unique_ptr<Impl>(new Impl(params));
+    }
+        
+    BallLogLikelihood::~BallLogLikelihood() = default;
+
+    void BallLogLikelihood::operator()(const cv::cuda::GpuMat& src, cv::cuda::GpuMat& dst, cv::cuda::Stream& stream) {
+      return _impl->bll(src, dst, stream);
+    }
+
+    class Binarizer::Impl {
+      public:
+        double thresh;
+        GpuMat bin_img;
+
+        Impl(const json& conf) {
+          if (conf.count("p_thresh")) {
+            double p = conf.at("p_thresh");
+            thresh = log(p/(1-p)); //logit function
+          } else if (conf.count("thresh")) {
+            thresh = conf.at(thresh);
+          }
+        }
+
+        void bin(const cv::cuda::GpuMat& src, cv::cuda::GpuMat& dst, cv::cuda::Stream& stream) {
+          cv::cuda::threshold(src, bin_img, thresh, 1.0, THRESH_BINARY_INV, stream);
+          dst = GpuMat(bin_img.rows, bin_img.cols, CV_8UC1);
+          bin_img.convertTo(dst, CV_8UC1, 250.0, 0.0, stream);
+        }
+    };
+
+    Binarizer::Binarizer(const nlohmann::json& params) {
+      _impl = unique_ptr<Impl>(new Impl(params));
+    }
+
+    Binarizer::~Binarizer() = default;
+
+    void Binarizer::operator()(const cv::cuda::GpuMat& src, cv::cuda::GpuMat& dst, cv::cuda::Stream& stream) {
+      _impl->bin(src, dst, stream);
+    }
+
+
+
+  };
+
+};
+