edit allover project

Makefile

 all:
-	nvcc src/radix_sort_cuda.cu -o radix_sort_cuda
+	nvcc src/radix_sort_par.cu -o radix_sort_cuda
 
 clean:
 	rm -rf radix_sort_cuda

src/radix_sort_par.cu

@@ -5,24 +5,30 @@
 using namespace std; 
   
 // A utility function to get maximum value in arr[] 
-__device__ int getMax(int arr[], int n) 
+int getMax(int arr[], int n) 
 { 
     int mx = arr[0]; 
     for (int i = 1; i < n; i++) 
         if (arr[i] > mx) 
             mx = arr[i]; 
     return mx; 
+}
+
+__global__ void copyArray(int *arr, int *output, int n)
+{
+    int id = blockIdx.x*blockDim.x+threadIdx.x;
+
+    if (id<n)
+        arr[id]=output[id];
 } 
   
 // A function to do counting sort of arr[] according to 
 // the digit represented by exp. 
-__device__ void countSort(int arr[], int n, int exp) 
+void countSort(int arr[], int n, int exp) 
 { 
-//	int index = threadIdx.x;
-//    int stride = blockDim.x;
-
-    int *output= (int*)malloc(sizeof(int)*n); // output array 
-    int i, count[10] = {0}; 
+    int output[n]; // output array 
+    long i;
+    int count[10] = {0}; 
   
     // Store count of occurrences in count[] 
     for (i = 0; i < n; i++) 
@@ -42,42 +48,59 @@ __device__ void countSort(int arr[], int n, int exp)
   
     // Copy the output array to arr[], so that arr[] now 
     // contains sorted numbers according to current digit 
-    for (i = 0; i < n; i++) 
-        arr[i] = output[i]; 
+    //for (i = 0; i < n; i++) 
+        //arr[i] = output[i];
+
+    int *h_arr;
+    int *h_output;
+    int *d_arr;
+    int *d_output;
+
+    size_t bytes= n*sizeof(int);
+
+    cudaMalloc(&d_arr, bytes);
+    cudaMalloc(&d_output, bytes);
+    
+    if (d_output==0){
+        printf ("hi");
+    }
+
+    cudaMemcpy(d_output,output, bytes, cudaMemcpyHostToDevice);
+
+    int blockSize, gridSize;
+  
+    // Number of threads in each thread block
+    blockSize = 256;
+      
+    // Number of thread blocks in grid
+    gridSize = ceil((float)n/blockSize);
+
+    copyArray<<<gridSize,blockSize>>>(d_arr,d_output,n); 
+   
+    cudaMemcpy(arr,d_arr,bytes,cudaMemcpyDeviceToHost);
+    
+    cudaFree(d_arr);
+    cudaFree(d_output);
+
 } 
   
 // The main function to that sorts arr[] of size n using  
-// Radix Sort
-
-__global__ void radixsort(int *arr, int n) 
+// Radix Sort 
+void radixsort(int arr[], int n) 
 { 
-    //int *d_arr;
-
     // Find the maximum number to know number of digits 
     int m = getMax(arr, n); 
   
     // Do counting sort for every digit. Note that instead 
     // of passing digit number, exp is passed. exp is 10^i 
     // where i is current digit number 
-
-    // allocate device memory
-    //cudaMalloc((void**)&d_arr,sizeof(int)*n);
-
-    //cudaMemcpy(d_arr, arr, sizeof(int)*n,cudaMemcpyHostToDevice);
     for (int exp = 1; m/exp > 0; exp *= 10) 
-        //countSort<<<1,1024>>>(d_arr, n, exp); 
-        countSort(arr,n,exp);
-        //transfer data back to host memory
-        //cudaMemcpy(arr, d_arr, sizeof(int)*n, cudaMemcpyDeviceToHost);
-        
-    //deallocate device memory
-    //cudaFree(d_arr);
-    __syncthreads();
-
+        countSort(arr, n, exp); 
 } 
   
 // A utility function to print an array 
-void print(int arr[], int n){ 
+void print(int arr[], int n) 
+{ 
     for (int i = 0; i < n; i++) 
         cout << arr[i] << "\n"; 
 } 
@@ -109,18 +132,12 @@ int main(int argc, char *argv[])
 { 
     timespec start, stop;
     
-    int *d_arr;
     int n;
     n= atoi(argv[1]);
     int arr[n];
     rng(arr,n);
-
-    cudaMalloc((void**)&d_arr,sizeof(int)*n);
-    cudaMemcpy(d_arr, arr, sizeof(int)*n,cudaMemcpyHostToDevice);
-
     clock_gettime(CLOCK_REALTIME, &start);
-    radixsort<<<1,32>>>(d_arr,n);
-    cudaMemcpy(arr, d_arr, sizeof(int)*n,cudaMemcpyDeviceToHost);
+    radixsort(arr,n);
     clock_gettime(CLOCK_REALTIME, &stop);
     print(arr,n);
     
@@ -128,9 +145,5 @@ int main(int argc, char *argv[])
     long time = duration.tv_sec * 1000000 + duration.tv_nsec/1000;
     printf("\n%d.%09d s\n", duration.tv_sec, duration.tv_nsec);
     
-    //deallocate host memory
-    cudaFree(d_arr);
-    
     return 0; 
 } 
-