cudagpgputhrustcub

May I use CUDA CUB iterator instead of thrust?


Is it possible to use iterators with CUB like Thrust? I want t use CUB instead of thrust as follow:

__global__ void reduce_roster(thrust::device_vector<float>::iterator vect, float * tab_seq, int SEUIL_ROSTER)
{
    int tid = blockIdx.x * blockDim.x + threadIdx.x;
    float resultat = 0;

    int i = TAILLE_ROSTER/TAILLE_SEQ_ROSTER;
    resultat = thrust::reduce(thrust::device, vect + (tid * TAILLE_ROSTER) + TAILLE_SEQ_ROSTER *  (i - 1), vect + (tid * TAILLE_ROSTER) + TAILLE_SEQ_ROSTER * i);


    float tmp;
    i--;

    while (resultat != -1 && i != 0)
    {
        if (resultat > SEUIL_ROSTER)
        {
            resultat = -1; 
        }
        else
        {               

            tmp = thrust::reduce(thrust::device, vect + (tid * TAILLE_ROSTER) + TAILLE_SEQ_ROSTER *  (i - 1), vect + (tid * TAILLE_ROSTER) + TAILLE_SEQ_ROSTER * i);

            resultat = resultat + tmp;


            i--;
        }
    }

Since CUB is faster than trust I'm trying to use it to reduce segments of an array.


Solution

  • No.

    And further to that, it isn't possible to use CUB to perform the same operation as you have shown in your code either. Depending on how you compile your kernel, thrust will either emit code where the reduction is performed serially by each thread, or by a child kernel launch using dynamic parallelism. CUB, on the other hand, has warp and block scope device reduction functions. It does not support single thread or dynamic parallelism scope operations.