GetPointer

struct GpuMat_PointerData
{
    uintptr_t cudaPtr;
    int64_t   step;
    uint64_t  totalBytes;
    int32_t   cols;
    int32_t   rows;
    int32_t   elemSize;
    int32_t   cvType;
    int32_t   channels;
    uintptr_t handlePtr;
    uintptr_t releaseCallbackPtr;
};

typedef void (*ReleasePointerCallback)(void* handle);


//  CUDA Kernel
__global__ void IncrementByOneKernel(uint8_t* data, int step, int cols, int rows)
{
    // Derive the 2-D pixel coordinate from the launch grid.
    const int x = blockIdx.x * blockDim.x + threadIdx.x;
    const int y = blockIdx.y * blockDim.y + threadIdx.y;

    // Discard threads that map outside the image boundaries.
    if (x >= cols || y >= rows)
    {
        return;
    }

    // Compute the linear byte offset using the row stride to account for
    // any alignment padding introduced by the CUDA memory allocator.
    const int index = (y * step) + x;

    // Increment the pixel value
    data[index] += 1;
}

extern "C" __declspec(dllexport) void ProcessGpuDataWithKernel(GpuMat_PointerData* data)
{
    // Validate the descriptor and device address before dereferencing either.
    if (data == nullptr || data->cudaPtr == 0)
    {
        return;
    }

    uint8_t*               deviceBuffer = reinterpret_cast<uint8_t*>(data->cudaPtr);
    void*                  handle       = reinterpret_cast<void*>(data->handlePtr);
    ReleasePointerCallback releaseMem   = reinterpret_cast<ReleasePointerCallback>(data->releaseCallbackPtr);

    // Process only single-channel 8-bit images (CV_8UC1, OpenCV type value 0).
    if (data->cvType == 0)
    {
        // Partition the image into 16×16 thread blocks and compute the minimum
        // number of blocks required to cover every pixel in the image.
        const dim3 threadsPerBlock(16, 16);
        const dim3 numBlocks(
            (data->cols + threadsPerBlock.x - 1) / threadsPerBlock.x,
            (data->rows + threadsPerBlock.y - 1) / threadsPerBlock.y
        );

        // Dispatch the kernel asynchronously on the default CUDA stream.
        IncrementByOneKernel<<<numBlocks, threadsPerBlock>>>(
            deviceBuffer,
            static_cast<int>(data->step),
            data->cols,
            data->rows
        );

        // Synchronise the host thread with the device to ensure the kernel has
        // completed before control returns to LabVIEW or cleanup proceeds.
        cudaDeviceSynchronize();
    }

    // Invoke the OpenCV DLL release callback to transfer buffer ownership back
    // to the runtime, allowing it to decrement the reference count and free the handle.
    if (handle != nullptr && releaseMem != nullptr)
    {
        releaseMem(handle);
    }
}