And here it is already interesting, the threadIdx variable appeared, which we did not seem to declare anywhere. Yes, the system provides us with it. Imagine that in the previous example, the array has three elements, and you want to run it on three parallel threads. To do this, you would need to add one more parameter – the index or number of the stream. This is what the video card does for us, although it passes the index as a static variable and can work with several dimensions at once – x, y, z.

Another nuance, if you are going to run a large number of parallel threads at once, then the threads will have to be divided into blocks (an architectural feature of video cards). The maximum block size depends on the video card, and the index of the element for which we are performing calculations will need to be obtained as follows:

int i = blockIdx.x * blockDim.x + threadIdx.x; // blockIdx – block index, blockDim – block size, threadIdx – thread index in the block

As a result, what we have: a set of parallel working threads executing the same code, but with different indices, and, accordingly, data, i.e. the same SIMD.

This is the simplest example, but if you want to work with a GPU, your task must be reduced to the same form. Unfortunately, this is not always possible and in some cases may become the topic of a doctoral dissertation, but nevertheless, classical algorithms can still be reduced to this form.