#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 = 5, quiet = 0, timing = 0, platform = -1, device = -1; float lat = 30, lng = 90; // parse command line if (parseCommandline(argc, argv, filename, &resultsCount, &lat, &lng, &quiet, &timing, &platform, &device)) { printUsage(); return 0; } int numRecords = loadData(filename, records, locations); // for(i=0;i numRecords) resultsCount = numRecords; context = cl_init_context(platform, device, quiet); 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 < resultsCount; i++) { printf("%s --> 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 **/ int q = 0; 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) { printf("error reading filelist\n"); exit(0); } printf("loading db: %s\n", dbname); 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++; if (0 == (recNum % 500)) break; } if (++q == 3) break; 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 < topN; i++) { minLoc = i; for (j = i; j < numRecords; j++) { val = distances[j]; if (val < distances[minLoc]) minLoc = j; } // swap locations and distances tempRec = &records[i]; records[i] = records[minLoc]; records[minLoc] = *tempRec; tempDist = distances[i]; distances[i] = distances[minLoc]; distances[minLoc] = tempDist; // add distance to the min we just found records[i].distance = distances[i]; } } int parseCommandline(int argc, char *argv[], char *filename, int *r, float *lat, float *lng, int *q, int *t, int *p, int *d) { int i; // if (argc < 2) return 1; // error strncpy(filename, "filelist.txt", 100); char flag; for (i = 1; i < argc; i++) { if (argv[i][0] == '-') { // flag flag = argv[i][1]; switch (flag) { case 'r': // number of results i++; *r = atoi(argv[i]); break; case 'l': // lat or lng if (argv[i][2] == 'a') { // lat *lat = atof(argv[i + 1]); } else { // lng *lng = atof(argv[i + 1]); } i++; break; case 'h': // help return 1; break; case 'q': // quiet *q = 1; break; case 't': // timing *t = 1; break; case 'p': // platform i++; *p = atoi(argv[i]); break; case 'd': // device i++; *d = atoi(argv[i]); break; } } } if ((*d >= 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("\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