Возврат с GPU uchar и преобразование его в Mat, не получается сделать корректные вычисления

Суть задачи проста: применить к заданному каналу изображения операцию "негатив".
У меня есть программа, которая решает поставленную задачу просто на cpp с использованием opencv (cpp + opencv исходные требования):

int main(int argc, char** argv)
{
    // Чтение файла
    Mat image = imread("in_img.jpg");
    if (!image.data)    
    {
        cout << "Could not open or find the image" << std::endl;
        return -1;
    }

    // Работа с изображением
    int channel = 2;                                // изменяемый канал
    for (int i = 0; i < image.rows; i++)
    {
        Vec3b* p = image.ptr<Vec3b>(i);             //указатель на 1-й пиксель в строке
        
        for (int j = 0; j < image.cols; j++)
            p[j][channel] = 255 - p[j][channel];    // применяем к заданному каналу операцию "негатив"
    }

    // Показ изображения
    namedWindow("Display window", WINDOW_AUTOSIZE); 
    imshow("Display window", image);                
    waitKey(0);                                     
    return 0;
}

Она работает.
Теперь мне необходимо переложить все на GPU (Cuda). Я все настроил, все запускается, изучил базу по cuda, но с выполнением задания получается косяк. Изображение итоговое просто черное. Не могу понять в чем проблема.
Вот код (файл .cu):

#include <opencv2/core/core.hpp>
#include <opencv2/highgui/highgui.hpp>
#include <iostream>

#include <stdio.h>
#include <string.h>
#include <memory.h>

#include "cuda_runtime.h"
#include "device_launch_parameters.h"

using namespace cv;
using namespace std;

#pragma region Проверка на ошибки

static void Check(cudaError_t err, const char* file, int line) {
    if (err != cudaSuccess) {
        printf("Error: %s at line %d in file %s\n", cudaGetErrorString(err), line, file);
        exit(EXIT_FAILURE);
    }
}
#define CHECK( err ) (Check( err, __FILE__, __LINE__ ))

#pragma endregion 

#pragma region Макрос для Intellisense

#ifdef __CUDACC__
#define KERNEL_ARGS2(grid, block) <<< grid, block >>>
#define KERNEL_ARGS3(grid, block, sh_mem) <<< grid, block, sh_mem >>>
#define KERNEL_ARGS4(grid, block, sh_mem, stream) <<< grid, block, sh_mem, stream >>>
#else
#define KERNEL_ARGS2(grid, block)
#define KERNEL_ARGS3(grid, block, sh_mem)
#define KERNEL_ARGS4(grid, block, sh_mem, stream)
#endif

#pragma endregion


// GPU: применение к заданному каналу изображения операции "негатив"
__global__ void negative(uchar* img, int channel, int N, uchar* out)
{
    int i = 3 * (blockIdx.x * blockDim.x + threadIdx.x);
    if (i >= N)
        return;
    out[i + channel] = 255 - img[i + channel];
}

int Launch()
{
    // Чтение файла
    Mat image = imread("in_img.jpg");
    if (!image.data)    // проверка корректного входа
    {
        cout << "Could not open or find the image" << std::endl;
        return -1;
    }
    int _width = image.cols;
    int _height = image.rows;
    int _type = image.type();

    int channel = 2;    // изменяемый канал

    uchar* host_img = image.data;

    size_t N = image.rows * image.cols * 3;

    uchar* dev_img;
    uchar* dev_imgOut;

    CHECK(cudaMalloc((void**)&dev_img, N * sizeof(uchar)));
    CHECK(cudaMalloc((void**)&dev_imgOut, N * sizeof(uchar)));

    CHECK(cudaMemcpy(dev_img, host_img, N * sizeof(uchar), cudaMemcpyHostToDevice));

    negative KERNEL_ARGS2(((N + 511) / 512), 512) (dev_img, channel, N, dev_imgOut);

    CHECK(cudaMemcpy(host_img, dev_imgOut, N * sizeof(uchar), cudaMemcpyDeviceToHost));

    Mat imageOut = Mat(_height, _width, _type, host_img);

    CHECK(cudaFree(dev_img));
    CHECK(cudaFree(dev_imgOut));

    // Показ изображения
    namedWindow("Display window", WINDOW_AUTOSIZE); // Создания окна отображения
    imshow("Display window", imageOut);             // Внутри окна показываем изображение
    waitKey(0);                                     // Ожидание нажатия клавиши в окне
    
    return 0;
}

Есть еще файл .cpp, но в нем просто main, в котором вызывается .cu.
"Почему сразу все не в .cu?" -> Я работаю через windows 10, visual studio 2019. Просто в .cu подрубить opencv не получилось, обхожу через проект .cpp.
Еще раз суть проблемы: итоговое imageOut получается просто черным. Я предполагаю, что я не верно преобразую uchar в Mat. Помогите, пожалуйста, разобраться в чем я накосячил.
p.s.: первоначально код ядра был такой:

__global__ void negative(uchar* img, int channel, int N)
{
    int i = 3 * (blockIdx.x * blockDim.x + threadIdx.x);
    if (i >= N)
        return;
    img[i + channel] = 255 - img[i + channel];
}

остальное было тем же самым, но итоговое изображение по факту вывода было точной копией входного без изменения