Separazione numeri pari e dispari in CUDA

Question

mi hanno una serie di numeri come {1,2,3,4,5,6,7,8,9,10} e voglio separare anche e numeri dispari come:

even = {2,4,6,8} 

e:

odd = {1,3,5,7} 

Sono consapevole delle atomica operazioni in CUDA, e anche consapevoli del fatto che l'uscita non dovrebbe soffrire di condizioni di gara. Non voglio usare operazioni atomiche. Come posso ottenere questo risultato senza utilizzare parole chiave atomiche?

CODICE:

#include <stdio.h> 
#include <cuda.h> 

// Kernel that executes on the CUDA device 
__global__ void square_array(float *total,float *even,float *odd, int N) 
{ 
    int idx = blockIdx.x * blockDim.x + threadIdx.x; 
    int a=total[idx]; 
    if ((a%2)==0) 
    { 
    for (int i=0;i<=idx;i++) 
    { 
     int b = even[i]; 
     if(b==0) 
     { 
      even[i] = total[idx]; 
      break; 

     } 
    } 
    } 
    else 
     { 
    for (int i=0;i<idx;i++) 
    { 
     int c = odd[i]; 

      odd[i] = total[idx]; 
      break; 
    } 
    } 
} 

// main routine that executes on the host 
int main(void) 
{ 
    float *total_h,*even_h, *odd_h,*total_d, *even_d,*odd_d; // Pointer to host & device arrays 
    const int N = 10; // Number of elements in arrays 
    size_t size = N * sizeof(float); 


    total_h = (float *)malloc(size); // Allocate array on host 
    even_h = (float *)malloc(size); // Allocate array on host 
    odd_h = (float *)malloc(size); // Allocate array on host 

    cudaMalloc((void **) &total_d, size); 
    cudaMalloc((void **) &even_d, size); 
    cudaMemset(even_d,0,size);   
    cudaMalloc((void **) &odd_d, size); // Allocate array on device 
    cudaMemset(odd_d,0,size); 


    // Initialize host array and copy it to CUDA device 
    for (int i=0; i<N; i++) total_h[i] = (float)i+1; 
    cudaMemcpy(total_d, total_h, size, cudaMemcpyHostToDevice); 
    // Do calculation on device: 

    square_array <<< 1,10 >>> (total_d,even_d,odd_d, N); 
    // Retrieve result from device and store it in host array 

    cudaMemcpy(even_h, even_d, sizeof(float)*N, cudaMemcpyDeviceToHost); 
    cudaMemcpy(odd_h, odd_d, sizeof(float)*N, cudaMemcpyDeviceToHost); 

    // Print results 
    printf("total Numbers\n"); 
    for (int i=0; i<N; i++) printf("%f\n",total_h[i]); 

    printf("EVEN Numbers\n"); 
    for (int i=0; i<N; i++) printf("%f\n",even_h[i]); 

    printf("ODD Numbers\n"); 
    for (int i=0; i<N; i++) printf("%f\n",odd_h[i]); 
    // Cleanup 
    free(total_h); 
    free(even_h); 
    free(odd_h); 


    cudaFree(total_d); 
    cudaFree(even_d); 
    cudaFree(odd_d); 
} 

USCITA:

Answer 1

Come suggerito da Jared Hoberock, sarebbe molto più facile da utilizzare l'algoritmo di partizionamento efficiente disponibile in CUDA di spinta invece di iniziare lo sviluppo di una routine di partizionamento personalizzata. Di seguito, si prega di trovare un esempio completo funzionante.

#include <thrust\device_vector.h> 
#include <thrust\partition.h> 
#include <thrust\execution_policy.h> 

struct is_even { __host__ __device__ bool operator()(const int &x) { return (x % 2) == 0; } }; 

void main() { 

    const int N = 10; 

    thrust::host_vector<int> h_data(N); 
    for (int i=0; i<N; i++) h_data[i] = i; 

    thrust::device_vector<int> d_data(h_data); 
    thrust::device_vector<int> d_evens(N/2); 
    thrust::device_vector<int> d_odds(N/2); 

    thrust::partition_copy(d_data.begin(), d_data.end(), d_evens.begin(), d_odds.begin(), is_even()); 

    printf("Even numbers\n"); 
    for (int i=0; i<N/2; i++) { 
     int val = d_evens[i]; 
     printf("evens[%i] = %i\n",i,val); 
    } 

    printf("Odd numbers\n"); 
    for (int i=0; i<N/2; i++) { 
     int val = d_odds[i]; 
     printf("odds[%i] = %i\n",i,val); 
    } 

}