Empty addresses returned by mem_trace

[dwz1212]

Hi, when I use the mem_trace to trace the test-apps vectoradd, I noticed that some memory traces contain empty addresses (0x0000000000000000). For example:

...... MEMTRACE: CTX 0x000055967fa50640 - grid_launch_id 0 - CTA 1,0,0 - warp 0 - LDG.E - 0x00007ff412a00890 0x00007ff412a00894 0x00007ff412a00898 0x00007ff412a0089c 0x00007ff412a00850 0x00007ff412a00854 0x00007ff412a00858 0x00007ff412a0085c 0x00007ff412a00860 0x00007ff412a00864 0x00007ff412a00868 0x00007ff412a0086c 0x00007ff412a00870 0x00007ff412a00874 0x00007ff412a00878 0x00007ff412a0087c 0x00007ff412a00880 0x00007ff412a00884 0x00007ff412a00888 0x00007ff412a0088c 0x00007ff412a00890 0x00007ff412a00894 0x00007ff412a00898 0x00007ff412a0089c 0x00007ff412a00850 0x00007ff412a00854 0x00007ff412a00858 0x00007ff412a0085c 0x00007ff412a00860 0x00007ff412a00864 0x00007ff412a00868 0x00007ff412a0086c MEMTRACE: CTX 0x000055967fa50640 - grid_launch_id 0 - CTA 1,0,0 - warp 4 - LDG.E - 0x00007ff412a00860 0x00007ff412a00864 0x00007ff412a00868 0x00007ff412a0086c 0x00007ff412a00870 0x00007ff412a00874 0x00007ff412a00878 0x00007ff412a0087c 0x00007ff412a00880 0x00007ff412a00884 0x00007ff412a00888 0x00007ff412a0088c 0x00007ff412a00890 0x00007ff412a00894 0x00007ff412a00898 0x00007ff412a0089c 0x0000000000000000 0x0000000000000000 0x0000000000000000 0x0000000000000000 0x0000000000000000 0x0000000000000000 0x0000000000000000 0x0000000000000000 0x0000000000000000 0x0000000000000000 0x0000000000000000 0x0000000000000000 0x0000000000000000 0x0000000000000000 0x0000000000000000 0x0000000000000000 MEMTRACE: CTX 0x000055967fa50640 - grid_launch_id 0 - CTA 0,0,0 - warp 6 - LDG.E - 0x00007ff412a00880 0x00007ff412a00884 0x00007ff412a00888 0x00007ff412a0088c 0x00007ff412a00890 0x00007ff412a00894 0x00007ff412a00898 0x00007ff412a0089c 0x00007ff412a00850 0x00007ff412a00854 0x00007ff412a00858 0x00007ff412a0085c 0x00007ff412a00860 0x00007ff412a00864 0x00007ff412a00868 0x00007ff412a0086c 0x00007ff412a00870 0x00007ff412a00874 0x00007ff412a00878 0x00007ff412a0087c 0x00007ff412a00880 0x00007ff412a00884 0x00007ff412a00888 0x00007ff412a0088c 0x00007ff412a00890 0x00007ff412a00894 0x00007ff412a00898 0x00007ff412a0089c 0x00007ff412a00850 0x00007ff412a00854 0x00007ff412a00858 0x00007ff412a0085c......

The vectoradd kernel and the grid/thread configuration are as follows:

__global__ void vecAdd(double *a, double *b, double *c, int n) { int id = blockIdx.x * blockDim.x + threadIdx.x; if (id < n) c[id] = a[id] + b[id];} ...... int blockSize, gridSize; blockSize = 1024; gridSize = (int)ceil((float)n / blockSize); CUDA_SAFECALL((vecAdd<<<gridSize, blockSize>>>(d_a, d_b, d_c, n)));

I'm curious about the explanation for these empty addresses. Were these addresses actually accessed, or could they be related to some sort of alignment mechanism?

[christindbose]

hi, is this with the new nvbit tool?

[christindbose]

@dwz1212

[dwz1212]

hi, is this with the new nvbit tool?

Hi, it's with the version NVBit 1.7 but not 1.7.1.