#ifndef __NEAREST_NEIGHBOR__ #define __NEAREST_NEIGHBOR__ #include "nearestNeighbor.h" cl_context context=NULL; int main(int argc, char *argv[]) { std::vector records; float *recordDistances; //LatLong locations[REC_WINDOW]; std::vector locations; int i; // args char filename[100]; int resultsCount=10,quiet=0,timing=0,platform=-1,device=-1,use_gpu=-1; float lat=0.0,lng=0.0; // parse command line if (parseCommandline(argc, argv, filename,&resultsCount,&lat,&lng, &quiet, &timing, &platform, &device, &use_gpu)) { printUsage(); return 0; } int numRecords = loadData(filename,records,locations); //for(i=0;i numRecords) resultsCount = numRecords; context = cl_init_context(platform,device,quiet,use_gpu); recordDistances = OpenClFindNearestNeighbors(context,numRecords,locations,lat,lng,timing); // find the resultsCount least distances findLowest(records,recordDistances,numRecords,resultsCount); // print out results if (!quiet) for(i=0;i Distance=%f\n",records[i].recString,records[i].distance); } free(recordDistances); return 0; } float *OpenClFindNearestNeighbors( cl_context context, int numRecords, std::vector &locations,float lat,float lng, int timing) { // 1. set up kernel cl_kernel NN_kernel; cl_int status; cl_program cl_NN_program; cl_NN_program = cl_compileProgram( (char *)"nearestNeighbor_kernel.cl",NULL); NN_kernel = clCreateKernel( cl_NN_program, "NearestNeighbor", &status); status = cl_errChk(status, (char *)"Error Creating Nearest Neighbor kernel",true); if(status)exit(1); // 2. set up memory on device and send ipts data to device // copy ipts(1,2) to device // also need to alloate memory for the distancePoints cl_mem d_locations; cl_mem d_distances; cl_int error=0; d_locations = clCreateBuffer(context, CL_MEM_READ_ONLY, sizeof(LatLong) * numRecords, NULL, &error); d_distances = clCreateBuffer(context, CL_MEM_READ_WRITE, sizeof(float) * numRecords, NULL, &error); cl_command_queue command_queue = cl_getCommandQueue(); cl_event writeEvent,kernelEvent,readEvent; error = clEnqueueWriteBuffer(command_queue, d_locations, 1, // change to 0 for nonblocking write 0, // offset sizeof(LatLong) * numRecords, &locations[0], 0, NULL, &writeEvent); // 3. send arguments to device cl_int argchk; argchk = clSetKernelArg(NN_kernel, 0, sizeof(cl_mem), (void *)&d_locations); argchk |= clSetKernelArg(NN_kernel, 1, sizeof(cl_mem), (void *)&d_distances); argchk |= clSetKernelArg(NN_kernel, 2, sizeof(int), (void *)&numRecords); argchk |= clSetKernelArg(NN_kernel, 3, sizeof(float), (void *)&lat); argchk |= clSetKernelArg(NN_kernel, 4, sizeof(float), (void *)&lng); cl_errChk(argchk,"ERROR in Setting Nearest Neighbor kernel args",true); // 4. enqueue kernel size_t globalWorkSize[1]; globalWorkSize[0] = numRecords; if (numRecords % 64) globalWorkSize[0] += 64 - (numRecords % 64); //printf("Global Work Size: %zu\n",globalWorkSize[0]); error = clEnqueueNDRangeKernel( command_queue, NN_kernel, 1, 0, globalWorkSize,NULL, 0, NULL, &kernelEvent); cl_errChk(error,"ERROR in Executing Kernel NearestNeighbor",true); // 5. transfer data off of device // create distances std::vector float *distances = (float *)malloc(sizeof(float) * numRecords); error = clEnqueueReadBuffer(command_queue, d_distances, 1, // change to 0 for nonblocking write 0, // offset sizeof(float) * numRecords, distances, 0, NULL, &readEvent); cl_errChk(error,"ERROR with clEnqueueReadBuffer",true); if (timing) { clFinish(command_queue); cl_ulong eventStart,eventEnd,totalTime=0; printf("# Records\tWrite(s) [size]\t\tKernel(s)\tRead(s) [size]\t\tTotal(s)\n"); printf("%d \t",numRecords); // Write Buffer error = clGetEventProfilingInfo(writeEvent,CL_PROFILING_COMMAND_START, sizeof(cl_ulong),&eventStart,NULL); cl_errChk(error,"ERROR in Event Profiling (Write Start)",true); error = clGetEventProfilingInfo(writeEvent,CL_PROFILING_COMMAND_END, sizeof(cl_ulong),&eventEnd,NULL); cl_errChk(error,"ERROR in Event Profiling (Write End)",true); printf("%f [%.2fMB]\t",(float)((eventEnd-eventStart)/1e9),(float)((sizeof(LatLong) * numRecords)/1e6)); totalTime += eventEnd-eventStart; // Kernel error = clGetEventProfilingInfo(kernelEvent,CL_PROFILING_COMMAND_START, sizeof(cl_ulong),&eventStart,NULL); cl_errChk(error,"ERROR in Event Profiling (Kernel Start)",true); error = clGetEventProfilingInfo(kernelEvent,CL_PROFILING_COMMAND_END, sizeof(cl_ulong),&eventEnd,NULL); cl_errChk(error,"ERROR in Event Profiling (Kernel End)",true); printf("%f\t",(float)((eventEnd-eventStart)/1e9)); totalTime += eventEnd-eventStart; // Read Buffer error = clGetEventProfilingInfo(readEvent,CL_PROFILING_COMMAND_START, sizeof(cl_ulong),&eventStart,NULL); cl_errChk(error,"ERROR in Event Profiling (Read Start)",true); error = clGetEventProfilingInfo(readEvent,CL_PROFILING_COMMAND_END, sizeof(cl_ulong),&eventEnd,NULL); cl_errChk(error,"ERROR in Event Profiling (Read End)",true); printf("%f [%.2fMB]\t",(float)((eventEnd-eventStart)/1e9),(float)((sizeof(float) * numRecords)/1e6)); totalTime += eventEnd-eventStart; printf("%f\n\n",(float)(totalTime/1e9)); } // 6. return finalized data and release buffers clReleaseMemObject(d_locations); clReleaseMemObject(d_distances); return distances; } int loadData(char *filename,std::vector &records,std::vector &locations){ FILE *flist,*fp; int i=0; char dbname[64]; int recNum=0; /**Main processing **/ flist = fopen(filename, "r"); while(!feof(flist)) { /** * Read in REC_WINDOW records of length REC_LENGTH * If this is the last file in the filelist, then done * else open next file to be read next iteration */ if(fscanf(flist, "%s\n", dbname) != 1) { fprintf(stderr, "error reading filelist\n"); exit(0); } fp = fopen(dbname, "r"); if(!fp) { printf("error opening a db\n"); exit(1); } // read each record while(!feof(fp)){ Record record; LatLong latLong; fgets(record.recString,49,fp); fgetc(fp); // newline if (feof(fp)) break; // parse for lat and long char substr[6]; for(i=0;i<5;i++) substr[i] = *(record.recString+i+28); substr[5] = '\0'; latLong.lat = atof(substr); for(i=0;i<5;i++) substr[i] = *(record.recString+i+33); substr[5] = '\0'; latLong.lng = atof(substr); locations.push_back(latLong); records.push_back(record); recNum++; } fclose(fp); } fclose(flist); return recNum; } void findLowest(std::vector &records,float *distances,int numRecords,int topN){ int i,j; float val; int minLoc; Record *tempRec; float tempDist; for(i=0;i= 0 && *p<0) || (*p>=0 && *d<0)) // both p and d must be specified if either are specified return 1; return 0; } void printUsage(){ printf("Nearest Neighbor Usage\n"); printf("\n"); printf("nearestNeighbor [filename] -r [int] -lat [float] -lng [float] [-hqt] [-p [int] -d [int]]\n"); printf("\n"); printf("example:\n"); printf("$ ./nearestNeighbor filelist.txt -r 5 -lat 30 -lng 90\n"); printf("\n"); printf("filename the filename that lists the data input files\n"); printf("-r [int] the number of records to return (default: 10)\n"); printf("-lat [float] the latitude for nearest neighbors (default: 0)\n"); printf("-lng [float] the longitude for nearest neighbors (default: 0)\n"); printf("\n"); printf("-h, --help Display the help file\n"); printf("-q Quiet mode. Suppress all text output.\n"); printf("-t Print timing information.\n"); printf("\n"); printf("-p [int] Choose the platform (must choose both platform and device)\n"); printf("-d [int] Choose the device (must choose both platform and device)\n"); printf("-g [int] Choose 1 for GPU, 0 for CPU\n"); printf("\n"); printf("\n"); printf("Notes: 1. The filename is required as the first parameter.\n"); printf(" 2. If you declare either the device or the platform,\n"); printf(" you must declare both.\n\n"); } #endif