unpackTest.cpp

// Copyright (c) 2012-2017 VideoStitch SAS
// Copyright (c) 2018 stitchEm
//
// Basic input unpacking tests.

#include "gpu/testing.hpp"
#include "gpu/util.hpp"

#include "libvideostitch/context.hpp"
#include "gpu/allocator.hpp"
#include "gpu/2dBuffer.hpp"

#include <image/unpack.hpp>
#include "libvideostitch/profile.hpp"
#include "libvideostitch/logging.hpp"

// Uncomment this to compare with opencv. You need to have OpenCV buit with GPU support, and add opencv_gpu to your
// libs. #define WITH_OPENCV_CMP

#ifdef WITH_OPENCV_CMP
#include <opencv2/core/core.hpp>
#include <opencv2/gpu/gpu.hpp>
#include <opencv2/opencv.hpp>
#endif

#include <memory>

#ifdef DUMP_IMAGES
#include "util/debugUtils.hpp"
#endif

namespace {
struct NoProfiler {
  NoProfiler(const char*, bool, VideoStitch::ThreadSafeOstream&) {}
};
}  // namespace

using namespace VideoStitch::Image;

namespace VideoStitch {
namespace Testing {

template <class ProfilerT>
void testOutOfPlaceRGB(unsigned width, unsigned height) {
  // Generate data
  std::vector<unsigned char> input;
  for (unsigned i = 0; i < width * height; ++i) {
    input.push_back((unsigned char)((33 * i) % 235));
    input.push_back((unsigned char)((31 * i) % 83));
    input.push_back((unsigned char)((42 * i) % 57));
  }
  DeviceBuffer<uchar3> inputBuffer(width, height);
  inputBuffer.fillData(input);

  auto potUniqueStream = GPU::UniqueStream::create();
  ENSURE(potUniqueStream.ok());
  auto stream = potUniqueStream.ref().borrow();

  DeviceBuffer<uint32_t> rgb210Buffer(width, height);
  {
    ENSURE(stream.synchronize());
    ProfilerT p("RGB pack", false, Logger::get(Logger::Info));
    // packUnpack(PixelArray<Image::RGB210Pixel>(width, height, rgb210Buffer.gpuBuf()),
    // PixelArray<Image::ConstRGBSolidPixel>(width, height, inputBuffer.gpuBuf()), stream);
    ENSURE(stream.synchronize());
  }

  DeviceBuffer<uchar3> outputBuffer(width, height);
  {
    ENSURE(stream.synchronize());
    ProfilerT p("RGB unpack", false, Logger::get(Logger::Info));
    // packUnpack(PixelArray<Image::RGBSolidPixel>(width, height, outputBuffer.gpuBuf()),
    // PixelArray<Image::ConstRGB210Pixel>(width, height, rgb210Buffer.gpuBuf()), stream);
    ENSURE(stream.synchronize());
  }

  std::vector<unsigned char> output;
  outputBuffer.readbackData(output);

  for (unsigned i = 0; i < width * height; ++i) {
    ENSURE_EQ((int)input[3 * i + 0], (int)output[3 * i + 0]);
    ENSURE_EQ((int)input[3 * i + 1], (int)output[3 * i + 1]);
    ENSURE_EQ((int)input[3 * i + 2], (int)output[3 * i + 2]);
  }
}

template <class ProfilerT>
void testYV12Minimal(unsigned width, unsigned height) {
  ENSURE(!(width & 1), "width must be even");
  ENSURE(!(height & 1), "height must be even");
  // Generate YV12 data
  unsigned char* input = new unsigned char[3 * width * height / 2];
  for (unsigned i = 0; i < width * height; ++i) {
    input[i] = 124;
  }
  for (unsigned i = 0; i < width * height / 4; ++i) {
    input[width * height + i] = 94;
  }
  for (unsigned i = 0; i < width * height / 4; ++i) {
    input[width * height + width * height / 4 + i] = 196;
  }

  auto buffer = GPU::uniqueBuffer<unsigned char>(width * height * 3 / 2, "unpackTest");
  ENSURE(buffer.status(), "no device memory");
  ENSURE(GPU::memcpyBlocking(buffer.borrow(), input), "transfer error");

  // to internal format
  auto surf = Core::OffscreenAllocator::createSourceSurface(width, height, "testYV12Minimal");
  ENSURE(surf.ok());
  convertYV12ToRGBA(*surf->pimpl->surface, buffer.borrow(), width, height, GPU::Stream::getDefault());

  // from internal format
  auto yOut = GPU::Buffer2D::allocate(width, height, "YV12 Y plane").value();
  auto uOut = GPU::Buffer2D::allocate(width / 2, height / 2, "YV12 U plane").value();
  auto vOut = GPU::Buffer2D::allocate(width / 2, height / 2, "YV12 V plane").value();
  unpackYV12(yOut, uOut, vOut, *surf->pimpl->surface, width, height, GPU::Stream::getDefault());

  // Check result of conversion
  std::vector<unsigned char> outputY(width * height);
  ENSURE(GPU::memcpyBlocking(outputY.data(), yOut), "transfer error");
  std::vector<unsigned char> outputU(width * height / 4);
  ENSURE(GPU::memcpyBlocking(outputU.data(), uOut), "transfer error");
  std::vector<unsigned char> outputV(width * height / 4);
  ENSURE(GPU::memcpyBlocking(outputV.data(), vOut), "transfer error");

  int eps = 1;
  for (unsigned i = 0; i < width * height; ++i) {
    ENSURE_APPROX_EQ(124, (int)outputY[i], eps);
  }
  for (unsigned i = 0; i < (width * height) / 4; ++i) {
    ENSURE_APPROX_EQ(94, (int)outputU[i], eps);
  }
  for (unsigned i = 0; i < (width * height) / 4; ++i) {
    ENSURE_APPROX_EQ(196, (int)outputV[i], eps);
  }

  // clean up
  delete[] input;
  ENSURE(yOut.release());
  ENSURE(uOut.release());
  ENSURE(vOut.release());
}

template <class ProfilerT>
void testNV12Minimal(unsigned width, unsigned height) {
  ENSURE(!(width & 1), "width must be even");
  ENSURE(!(height & 1), "height must be even");

  // Generate NV12 data
  std::vector<unsigned char> input(3 * width * height / 2);
  for (unsigned i = 0; i < width * height; ++i) {
    input[i] = 80 + i % 48;
  }
  for (unsigned i = 0; i < width * height / 2; i += 2) {
    input[width * height + i] = 90 + (i / width) % 80;
    input[width * height + i + 1] = 80 + (i / height) % 100;
  }

  // alloc
  auto buffer = GPU::uniqueBuffer<unsigned char>(width * height * 3 / 2, "unpackTest");
  ENSURE(buffer.status(), "no device memory");

  // transfer
  ENSURE(GPU::memcpyBlocking(buffer.borrow(), input.data()), "transfer error");

  // to internal format
  auto surf = Core::OffscreenAllocator::createSourceSurface(width, height, "testNV12Minimal");
  ENSURE(surf.ok());
  convertNV12ToRGBA(*surf->pimpl->surface, buffer.borrow(), width, height, GPU::Stream::getDefault());

#ifdef DUMP_IMAGES
  GPU::Stream::getDefault().synchronize();
  Debug::dumpRGBATexture("/tmp/nv12.png", *surf->pimpl->surface, width, height);
#endif

  // from internal format
  GPU::Buffer2D yOut = GPU::Buffer2D::allocate(width, height, "NV12 Y plane").value();
  GPU::Buffer2D uvOut = GPU::Buffer2D::allocate(width, height / 2, "NV12 UV plane").value();
  unpackNV12(yOut, uvOut, *surf->pimpl->surface, width, height, GPU::Stream::getDefault());

  // Check result of conversion
  std::vector<unsigned char> outputY(width * height);
  ENSURE(GPU::memcpyBlocking(outputY.data(), yOut), "transfer error");
  std::vector<unsigned char> outputUV(width * height / 2);
  ENSURE(GPU::memcpyBlocking(outputUV.data(), uvOut), "transfer error");

  GPU::Stream::getDefault().synchronize();

  int eps = 1;
  for (unsigned i = 0; i < width * height; ++i) {
    ENSURE_APPROX_EQ((int)(80 + i % 48), (int)outputY[i], eps);
  }
  for (unsigned i = 0; i < width * height / 2; i += 2) {
    ENSURE_APPROX_EQ((int)(90 + (i / width) % 80), (int)outputUV[i], eps);
    ENSURE_APPROX_EQ((int)(80 + (i / height) % 100), (int)outputUV[i + 1], eps);
  }

  // clean up
  ENSURE(yOut.release());
  ENSURE(uvOut.release());
}

// TODO_OPENCL_IMPL
#ifndef VS_OPENCL

void testBayerRGGB() {
  const int width = 6;
  const int height = 4;
  DeviceBuffer<unsigned char> bayerFiltered(width, height);
  const std::vector<unsigned char> inputData = {
      10, 42, 36, 78, 40, 52, 50, 64, 8, 82, 42, 44, 90, 112, 124, 122, 162, 56, 130, 144, 16, 48, 44, 128,
  };
  bayerFiltered.fill(inputData);

  PackedDeviceBuffer outputDev(width, height);
  convertBayerRGGBToRGBA(outputDev.gpuBuf(), bayerFiltered.gpuBufConst(), width, height, GPU::Stream::getDefault());

  std::vector<uint32_t> actual;
  outputDev.readback(actual);
  /*
  std::cout << "R" << std::endl;
  for (unsigned y = 0; y < height; ++y) {
    for (unsigned x = 0; x < width; ++x) {
      std::cout << RGBA::r(actual[y * width + x]) << " ";
    }
    std::cout << std::endl;
  }
  std::cout << "G" << std::endl;
  for (unsigned y = 0; y < height; ++y) {
    for (unsigned x = 0; x < width; ++x) {
      std::cout << RGBA::g(actual[y * width + x]) << " ";
    }
    std::cout << std::endl;
  }
  std::cout << "B" << std::endl;
  for (unsigned y = 0; y < height; ++y) {
    for (unsigned x = 0; x < width; ++x) {
      std::cout << RGBA::b(actual[y * width + x]) << " ";
    }
    std::cout << std::endl;
  }
  */

  // Red:
  //
  //   10,    ,  36,    ,  40,    ,                                                                      10,   23,   36,
  //   38,   40,   40, (40)
  //     ,    ,    ,    ,    ,    ,    =>                                                    =>          50,   65,   80,
  //     90,  101,  101, (XX)
  //   90,    , 124,    , 162,    ,                                                                      90,  107,  124,
  //   143,  162,  162, (162)
  //     ,    ,    ,    ,    ,    ,                                                                      90,  107,  124,
  //     143,  162,  162, (XX)
  //                                                                                                   (90), (XX),(124),
  //                                                                                                   (XX),(162), (XX),
  //                                                                                                   (162)

  // Green:
  //                                         (XX), (50), (XX),  (8), (XX), (42), (XX), (XX)      (XX), (50), (XX),  (8),
  //                                         (XX), (42), (XX), (XX)
  //     ,  42,    ,  78,    ,  52,          (42),     ,   42,     ,   78,     ,   52, (XX)      (42),   46,   42,   34,
  //     78,   53,   52, (XX)
  //   50,    ,   8,    ,  42,    ,     =>   (XX),   50,     ,    8,     ,   42,     , (42)  =>  (XX),   50,   53,    8,
  //   62,   42,   48, (42)
  //     , 112,    , 122,    ,  56,         (112),     ,  112,     ,  122,     ,   56, (XX)     (112),  101,  112,   64,
  //     122,   66,   56, (XX)
  //  130,    ,  16,    ,  44,    ,          (XX),  130,     ,   16,     ,   44,     , (44)      (XX),  130,   92,   16,
  //  76,   44,   50, (44)
  //                                         (XX), (XX),(112), (XX),(122), (XX), (56), (XX)      (XX), (XX),(112),
  //                                         (XX),(122), (XX), (56), (XX)

  // Blue:
  //                                                                                             (64), (XX), (64), (XX),
  //                                                                                             (82), (XX), (44),
  //     ,    ,    ,    ,    ,    ,                                                              (XX),   64,   64,   73,
  //     82,   63,   44, ,  64,    ,  82,    ,  44,    =>                                                    =>  (64),
  //     64,   64,   73,   82,   63,   44, ,    ,    ,    ,    ,    , (XX),  104,  104,   84,   65,   75,   86, , 144,
  //     ,  48,    , 128,                                                             (144),  144,  144,   96,   48, 88,
  //     128,

  const std::vector<uint32_t> expected = {
      RGBA::pack(10, 46, 64, 255),   RGBA::pack(23, 42, 64, 255),    RGBA::pack(36, 34, 73, 255),
      RGBA::pack(38, 78, 82, 255),   RGBA::pack(40, 53, 63, 255),    RGBA::pack(40, 52, 44, 255),
      RGBA::pack(50, 50, 64, 255),   RGBA::pack(65, 53, 64, 255),    RGBA::pack(80, 8, 73, 255),
      RGBA::pack(90, 62, 82, 255),   RGBA::pack(101, 42, 63, 255),   RGBA::pack(101, 48, 44, 255),
      RGBA::pack(90, 101, 104, 255), RGBA::pack(107, 112, 104, 255), RGBA::pack(124, 64, 84, 255),
      RGBA::pack(143, 122, 65, 255), RGBA::pack(162, 66, 75, 255),   RGBA::pack(162, 56, 86, 255),
      RGBA::pack(90, 130, 144, 255), RGBA::pack(107, 92, 144, 255),  RGBA::pack(124, 16, 96, 255),
      RGBA::pack(143, 76, 48, 255),  RGBA::pack(162, 44, 88, 255),   RGBA::pack(162, 50, 128, 255),
  };
  ENSURE_RGBA8888_ARRAY_EQ(expected.data(), actual.data(), width, height);
}

void benchBayerRGGB(const int width, const int height) {
  std::vector<unsigned char> inputData(width * height);
  for (int i = 0; i < width * height; ++i) {
    inputData[i] = (unsigned char)((i * 1357) % 255);
  }

  {
    DeviceBuffer<unsigned char> bayerFiltered(width, height);
    {
      Util::SimpleProfiler p("Debayer bench VS, copy to GPU", false, Logger::get(Logger::Info));
      bayerFiltered.fill(inputData);
    }
    PackedDeviceBuffer vsOutputDev(width, height);
    {
      Util::SimpleProfiler p("Debayer bench VS, debayer", false, Logger::get(Logger::Info));
      convertBayerRGGBToRGBA(vsOutputDev.gpuBuf(), bayerFiltered.gpuBufConst(), width, height,
                             GPU::Stream::getDefault());
      cudaDeviceSynchronize();
    }
    std::vector<uint32_t> vsOutput;
    {
      Util::SimpleProfiler p("Debayer bench VS, readback", false, Logger::get(Logger::Info));
      vsOutputDev.readback(vsOutput);
    }
  }

#ifdef WITH_OPENCV_CMP
  {
    std::unique_ptr<cv::gpu::GpuMat> cvSrc;
    {
      Util::SimpleProfiler p("Debayer bench OpenCV, create image, copy to GPU", false, std::cout);
      cvSrc.reset(new cv::gpu::GpuMat(cv::Mat(height, width, CV_8UC1, (void*)inputData.data())));
    }
    cv::gpu::GpuMat cvDevOutput(height, width, CV_8UC4);
    {
      Util::SimpleProfiler p("Debayer bench OpenCV, debayer", false, std::cout);
      cv::gpu::cvtColor(*cvSrc, cvDevOutput, CV_BayerBG2BGR);
      cudaDeviceSynchronize();
    }
    cv::Mat cvOutput(height, width, CV_8UC4);
    {
      Util::SimpleProfiler p("Debayer bench OpenCV, readback", false, std::cout);
      cvDevOutput.download(cvOutput);
      cudaDeviceSynchronize();
    }
  }
#endif
}

#ifdef WITH_OPENCV_CMP
void testBayerRGGBReal() {
  const int width = 600;
  const int height = 480;

  cv::Mat cvOutput(height, width, CV_8UC4);
  cv::Mat vsOutput(height, width, CV_8UC4);

  {
    DeviceBuffer<unsigned char> bayerFiltered(width, height);
    bayerFiltered.readRawFromFile("data/bayer/mire.gray");
    PackedDeviceBuffer vsOutputDev(width, height);
    {
      Util::SimpleProfiler p("Debayer real VS, debayer", false, std::cout);
      convertBayerRGGBToRGBA(vsOutputDev.ptr(), bayerFiltered.ptr(), width, height, 16);
      cudaDeviceSynchronize();
    }
    std::vector<uint32_t> tmp;
    vsOutputDev.readback(tmp);
    std::unique_ptr<std::ostream> os(Util::PpmWriter::openPpm("debayer-vs.ppm", width, height));
    for (int y = 0; y < height; ++y) {
      for (int x = 0; x < width; ++x) {
        const uint32_t v = tmp[y * width + x];
        *os << (char)RGBA::r(v);
        *os << (char)RGBA::g(v);
        *os << (char)RGBA::b(v);
      }
    }
    vsOutput = cv::Mat(height, width, CV_8UC4, (void*)tmp.data()).clone();
  }
  std::cout << vsOutput.at<uint32_t>(200, 200) << std::endl;

  {
    cv::gpu::GpuMat cvSrc(cv::imread("data/bayer/mire.pgm", CV_LOAD_IMAGE_GRAYSCALE));
    cv::gpu::GpuMat cvDevOutput(height, width, CV_8UC4);
    {
      Util::SimpleProfiler p("Debayer real OpenCV, debayer", false, std::cout);
      cv::gpu::cvtColor(cvSrc, cvDevOutput, CV_BayerBG2BGR);
      cudaDeviceSynchronize();
    }
    cvDevOutput.download(cvOutput);
  }
  cv::imwrite("debayer-opencv.ppm", cvOutput);
  // cv::imwrite("debayer-vs.ppm", vsOutput);
}
#endif

#endif  // VS_OPENCL

}  // namespace Testing
}  // namespace VideoStitch

int main(int argc, char** argv) {
  VideoStitch::Testing::initTest();
  VideoStitch::Testing::ENSURE(VideoStitch::GPU::setDefaultBackendDevice(0));

  // VideoStitch::Testing::testOutOfPlaceRGB<NoProfiler>(16, 16);
  // VideoStitch::Testing::testOutOfPlaceRGB<NoProfiler>(312, 531);

  // VideoStitch::Testing::testOutOfPlaceMonoY<NoProfiler>(16, 16);
  // VideoStitch::Testing::testOutOfPlaceMonoY<NoProfiler>(312, 531);

  VideoStitch::Testing::testYV12Minimal<NoProfiler>(16, 16);
  VideoStitch::Testing::testYV12Minimal<NoProfiler>(312, 532);

  VideoStitch::Testing::testNV12Minimal<NoProfiler>(16, 16);
  VideoStitch::Testing::testNV12Minimal<NoProfiler>(280, 124);

// TODO_OPENCL_IMPL
#ifndef VS_OPENCL

  // VideoStitch::Testing::testBayerRGGB();

#endif

#ifdef WITH_OPENCV_CMP
  VideoStitch::Testing::testBayerRGGBReal();
#endif

  if (argc > 1 && !strcmp(argv[1], "bench")) {
    VideoStitch::Testing::testOutOfPlaceRGB<VideoStitch::Util::SimpleProfiler>(6542, 5322);

// TODO_OPENCL_IMPL
#ifndef VS_OPENCL
    VideoStitch::Testing::benchBayerRGGB(6542, 5322);
#endif
  }
  VideoStitch::GPU::Context::destroy();
  return 0;
}