libargus UV plane layout in cuSurfObject?

Autonomous Machines Jetson & Embedded Systems Jetson TX2
1

From the argus/samples/cudaHistogram/main.cpp example,

I’m interested in expanding this to access the (YUV420) UV plane in CUDA. We see a surface created for the Y plane:

// Create a surface from the luminance plane
        CUDA_RESOURCE_DESC cudaResourceDesc;
        memset(&cudaResourceDesc, 0, sizeof(cudaResourceDesc));
        cudaResourceDesc.resType = CU_RESOURCE_TYPE_ARRAY;
        cudaResourceDesc.res.array.hArray = cudaEGLFrame.frame.pArray[0];
        CUsurfObject cudaSurfObj = 0;
        cuResult = cuSurfObjectCreate(&cudaSurfObj, &cudaResourceDesc);
        if (cuResult != CUDA_SUCCESS)
        {
            ORIGINATE_ERROR("Unable to create the surface object (CUresult %s)",
                getCudaErrorString(cuResult));
        }

and accessed via:

__global__ void histogram_smem_atomics(
    CUsurfObject surface,
    unsigned int width,
    unsigned int height,
    unsigned int *out)
{
...
surf2Dread(&data, surface, col, row);
...

We also see there is (what I assume is the UV plane) at cudaEGLFrame.frame.pArray[1] since there are 2 planes available.

The Y plane works fine, but I’m interested in knowing how the UV plane should be accessed within the 2D surface?
We create a seperate UV surface and pass it to CUDA in the same manner as the Y plane.

e.g. Is the U plane ‘width’ pixels wide by ‘height/4’ pixels high? Is it then followed by the V plane? or something else? In other words, what is the memory layout for the UV plane as a CUDA 2D surface (so we can read U and V values using surf2Dread with X,Y coordinates)?

2

Incase anyone else runs across this, I believe the UV plan is interleaved. The format is not YUV420 (Y followed by U followed by V) as the CUeglFrame colour format type suggests.

i.e.

Plane 0:
YYYYYYYY

Plane 1:
UVUV