Edit kernel.cu o complete the functionality of histogram， use a block size of 512.

here are three modes of operation for the application. Check main() for a description of the modes (repeated below). You will support each of these modes using Histogram with a privitization pattern.

kernel.cu

__global__ void histo_kernel(unsigned int* input, unsigned int* bins, unsigned int num_elements, unsigned int num_bins)
{

//insert code here

}

void histogram(unsigned int* input, unsigned int* bins, unsigned int num_elements, unsigned int num_bins) {

//insert code here

}

main.cu

#include
#include

#include "support.h"
#include "kernel.cu"

int main(int argc, char* argv[])
{
Timer timer;

// Initialize host variables ----------------------------------------------

printf("Setting up the problem..."); fflush(stdout);
startTime(&timer);

unsigned int *in_h;
unsigned int* bins_h;
unsigned int *in_d;
unsigned int* bins_d;
unsigned int num_elements, num_bins;
cudaError_t cuda_ret;

if(argc == 1) {
num_elements = 1000000;
num_bins = 4096;
} else if(argc == 2) {
num_elements = atoi(argv[1]);
num_bins = 4096;
} else if(argc == 3) {
num_elements = atoi(argv[1]);
num_bins = atoi(argv[2]);
} else {
printf(" Invalid input parameters!"
" Usage: ./histogram # Input: 1,000,000, Bins: 4,096"
" Usage: ./histogram # Input: m, Bins: 4,096"
" Usage: ./histogram # Input: m, Bins: n"
"");
exit(0);
}
initVector(&in_h, num_elements, num_bins);
bins_h = (unsigned int*) malloc(num_bins*sizeof(unsigned int));

stopTime(&timer); printf("%f s", elapsedTime(timer));
printf(" Input size = %u Number of bins = %u", num_elements,
num_bins);

// Allocate device variables ----------------------------------------------

printf("Allocating device variables..."); fflush(stdout);
startTime(&timer);

cuda_ret = cudaMalloc((void**)&in_d, num_elements * sizeof(unsigned int));
if(cuda_ret != cudaSuccess) printf("Unable to allocate device memory");
cuda_ret = cudaMalloc((void**)&bins_d, num_bins * sizeof(unsigned int));
if(cuda_ret != cudaSuccess) printf("Unable to allocate device memory");

cudaDeviceSynchronize();
stopTime(&timer); printf("%f s", elapsedTime(timer));

// Copy host variables to device ------------------------------------------

printf("Copying data from host to device..."); fflush(stdout);
startTime(&timer);

cuda_ret = cudaMemcpy(in_d, in_h, num_elements * sizeof(unsigned int),
cudaMemcpyHostToDevice);
if(cuda_ret != cudaSuccess) printf("Unable to copy memory to the device");

cuda_ret = cudaMemset(bins_d, 0, num_bins * sizeof(unsigned int));
if(cuda_ret != cudaSuccess) printf("Unable to set device memory");

cudaDeviceSynchronize();
stopTime(&timer); printf("%f s", elapsedTime(timer));

// Launch kernel ----------------------------------------------------------
printf("Launching kernel..."); fflush(stdout);
startTime(&timer);

histogram(in_d, bins_d, num_elements, num_bins);
cuda_ret = cudaDeviceSynchronize();
if(cuda_ret != cudaSuccess) printf("Unable to launch/execute kernel");

stopTime(&timer); printf("%f s", elapsedTime(timer));

// Copy device variables from host ----------------------------------------

printf("Copying data from device to host..."); fflush(stdout);
startTime(&timer);

cuda_ret = cudaMemcpy(bins_h, bins_d, num_bins * sizeof(unsigned int),
cudaMemcpyDeviceToHost);
if(cuda_ret != cudaSuccess) printf("Unable to copy memory to host");

cudaDeviceSynchronize();
stopTime(&timer); printf("%f s", elapsedTime(timer));

// Verify correctness -----------------------------------------------------

printf("Verifying results..."); fflush(stdout);

verify(in_h, bins_h, num_elements, num_bins);

// Free memory ------------------------------------------------------------

cudaFree(in_d); cudaFree(bins_d);
free(in_h); free(bins_h);

return 0;
}