CUDA. Адресация массива в ядре cuda и в потоке

Совсем недавно начал разбираться с CUDA и застрял на элементарном, не могу понять почему расходится адресация массива в ядре cuda и в потоке.

Создаю изображение, заполняю определенные пиксели:

Mat panoramicLeft(3840, 1404, CV_8UC3);

panoramicLeft.data[700 * 3840 + 1600] = 88;
panoramicLeft.data[700 * 3840 + 1600 + 1] = 89;
panoramicLeft.data[700 * 3840 + 1600 + 2] = 90;

panoramicLeft.data[750 * 3840 + 1700] = 77;
panoramicLeft.data[750 * 3840 + 1700 + 1] = 78;
panoramicLeft.data[750 * 3840 + 1700 + 2] = 79;

Копирую все это дело в память видеокарты:

cuda::GpuMat leftImagePanoramic.upload(panoramicLeft);

Есть простенькое ядро для проверки:

__global__ void gpuTest(uchar1* output, size_t outputWidth, size_t outputHeight, size_t offsetX, size_t offsetY)
{
    const int x = offsetX + blockIdx.x * blockDim.x + threadIdx.x;
    const int y = offsetY + blockIdx.y * blockDim.y + threadIdx.y;
    const int outputPosition = y * 3840 + x;
    if ((y == 700 && x == 1600) || (y == 750 && x == 1700))
    {
        uchar1 tmp1 = output[outputPosition];
        uchar1 tmp2 = output[outputPosition + 1];
        uchar1 tmp3 = output[outputPosition + 2];
        printf("CUDA DEBUG FF [%d, %d, %d, %d, %d]\n", (int)tmp1.x, (int)tmp2.x, tmp3.x, x, y);
    }
}


cudaError_t cudaTest(uchar1* output, size_t outputWidth, size_t outputHeight, size_t offsetX, size_t offsetY, cudaStream_t stream)
{
    if (offsetX < 0 || offsetY < 0 || outputWidth == 0 || outputHeight == 0)
    {
        return cudaErrorInvalidValue;
    }
    const dim3 blockDim(8, 8);
    const dim3 gridDim(iDivUp(outputWidth - offsetX, blockDim.x), iDivUp(outputHeight - offsetY, blockDim.y));
    gpuTest<<<gridDim, blockDim, 0, stream>>>(output, outputWidth, outputHeight, offsetX, offsetY);
    return cudaSuccess;
}

Запускаю его с полученным массивом:

cudaTest((uchar1*)leftImagePanoramic.data, 3840, 1404, 0, 0, 0);

Получаю что в указанных пикселях пусто:

CUDA DEBUG  [0, 0, 0, 1600, 700]
CUDA DEBUG  [0, 0, 0, 1700, 750]

Для проверки после того как ядро отработало, выгружаю массив обратно в оперативную память и проверяю:

leftImagePanoramic.download(panoramicLeft);
cout << "test after " << panoramicLeft.data[700 * 3840 + 1600] << " " << panoramicLeft.data[700 * 3840 + 1600 + 1] << " " << panoramicLeft.data[700 * 3840 + 1600 + 2];
cout << "test after " << panoramicLeft.data[750 * 3840 + 1700] << " " << panoramicLeft.data[750 * 3840 + 1700 + 1] << " " << panoramicLeft.data[750 * 3840 + 1700 + 2];

Получаю то, что и ожидалось:

test after 88 89 90
test after 77 78 79

Я наверняка в упор не вижу какой-то фигни, подскажите, пожалуйста...