Invalid results by summing an array, OpenCL, Interleaved Addressing

Question

i'm starting to learn OpenCl and as one of the tasks I have to write I program, that sums all elements of an array.

The program is supposed to be simple and I don't know what's wrong with me today, but it's not working. Well, it does, but sometimes it shows wrong results (sometimes doesn't).

The more elements we have, the bigger is the chance to get a wrong result (especially after 16536) The number of elements always equals to the power of two.

Could someone please tell me, what's wrong here?

The kernel:

__kernel void Reduction_InterleavedAddressing(__global uint* array, uint stride)
{
    unsigned int i = get_global_id (0); 
    unsigned int size = get_global_size(0);

    if ((i % stride*2) == 0 && (i + stride)<size){
            array[i]  += array[i+stride];
    }
}

Kernel call:

    unsigned int stride = 1;
    clErr = clSetKernelArg(m_InterleavedAddressingKernel, 0, sizeof(cl_mem), (void*)&m_dPingArray);

    for (; stride <= m_N / 2 ; stride*=2){
        clErr = clSetKernelArg(m_InterleavedAddressingKernel, 1, sizeof(cl_int), (void*)&stride);
        clErr = clEnqueueNDRangeKernel(CommandQueue, m_InterleavedAddressingKernel, 1, NULL, &globalWorkSize, LocalWorkSize, 0, NULL, NULL);
        V_RETURN_CL(clErr, "Error executing kernel");
    }

Thank your for your tips in advance

Answer 1

I testet it. You must set the stride * 2 to (stride * 2)! For my solution it worked well.

if ((i % (stride*2)) == 0 && (i + stride) < size) {
    array[i]  += array[i+stride];
}