Elementwise 2: The Same Add on the GPU

Lesson

In stage 1 one core walked every element in order. Here thousands of threads each handle one element, so the loop disappears and its counter becomes an index you compute from the thread's position in the grid.

Write three things:

1. This thread's global index, from `blockIdx.x`, `blockDim.x`, and `threadIdx.x`.
2. A bounds check. The grid is rounded up to cover N, so the last block has threads with no element to write. Without the guard they write past the end of the array.
3. The same assignment as stage 1.

After Run you get three separate times, and they answer different questions:

- **CPU loop time** — the host loop alone, the same measurement stage 1 reported.
- **GPU kernel time** — CUDA execution only, from CUDA events. Arithmetic, nothing else.
- **GPU end-to-end time** — copying inputs to the device, the kernel, copying the result back, and synchronizing. This is what a caller actually waits for.

Compare the CPU loop against end-to-end time when you want to know whether moving this operation to the GPU was worth it. Compare it against kernel time only when you want to know how fast the arithmetic itself is. Those are different questions and they can have different answers.

Many elements execute in parallel; the entire array does not reside or execute on the GPU simultaneously.

What you will learn

• Compute a global thread index from block and thread coordinates.
• Explain why a bounds check is required when the grid is rounded up to cover N.
• Distinguish GPU kernel time from GPU end-to-end time, and know which one a caller experiences.

Starter code

    // 1. Which element does this thread own? Build the index from the block
    //    it is in, how many threads each block has, and its position inside
    //    that block.
    // TODO: int i = ...

    // 2. The grid is rounded up to cover N, so the last block has spare
    //    threads. Without a guard they write past the end of C.
    // TODO: if (...) {

    // 3. The same line you wrote on the CPU.
    // TODO:     C[i] = A[i] + B[i];
    // TODO: }