CUDA Visionary Debugger

Current Step: 0x42A

Kernel Execution

// Sample CUDA kernel

__global__

void

vectorAdd

(

float

*A,

float

*B,

float

*C,

int

numElements) {

int

i = blockDim.x * blockIdx.x + threadIdx.x;

(i < numElements) {
C[i] = A[i] + B[i];
}
}

GPU Utilization

Execution Time

Memory Usage

Thread Visualization

PTX Assembly

.version 7.4

.target sm_86

.address_size 64

.visible .entry vectorAdd(

.param .u64 vectorAdd_param_0,

.param .u64 vectorAdd_param_1,

.param .u64 vectorAdd_param_2,

.param .u32 vectorAdd_param_3

)

.reg .pred %p<2>;

.reg .f32 %f<4>;

.reg .b32 %r<5>;

.reg .b64 %rd<7>;

ld.param.u64 %rd1, [vectorAdd_param_0];

ld.param.u64 %rd2, [vectorAdd_param_1];

ld.param.u64 %rd3, [vectorAdd_param_2];

ld.param.u32 %r1, [vectorAdd_param_3];

cvta.to.global.u64 %rd4, %rd1;

cvta.to.global.u64 %rd5, %rd2;

cvta.to.global.u64 %rd6, %rd3;

mov.u32 %r2, %tid.x;

mov.u32 %r3, %ctaid.x;

mov.u32 %r4, %ntid.x;

mad.lo.s32 %r5, %r3, %r4, %r2;

setp.lt.u32 %p1, %r5, %r1;

@%p1 bra LBB0_2;

LBB0_1:

ret;

LBB0_2:

mul.wide.s32 %rd7, %r5, 4;

add.s64 %rd8, %rd4, %rd7;

ld.global.f32 %f1, [%rd8];

add.s64 %rd9, %rd5, %rd7;

ld.global.f32 %f2, [%rd9];

add.f32 %f3, %f1, %f2;

add.s64 %rd10, %rd6, %rd7;

st.global.f32 [%rd10], %f3;

bra LBB0_1;

Register Values

Register	Type	Value	Thread