H_sum = new float[sumBufLen];

H_met = new float[metBufLen];

}

// call this before CalculateFFT()

void setup_clFFT()

{

/* FFT library realted declarations */

clfftDim dim = CLFFT_1D;

size_t clLengths[1] = { windowSize };

/* Setup clFFT. */

clfftSetupData fftSetup;

err = clfftInitSetupData(&fftSetup);

err = clfftSetup(&fftSetup);

/* Create a default plan for a complex FFT. */

err = clfftCreateDefaultPlan(&planHandle, ctx, dim, clLengths);

/* Set plan parameters. */

err = clfftSetPlanPrecision(planHandle, CLFFT_SINGLE);

err = clfftSetLayout(planHandle, CLFFT_REAL, CLFFT_HERMITIAN_INTERLEAVED);

err = clfftSetResultLocation(planHandle, CLFFT_OUTOFPLACE);

err = clfftSetPlanBatchSize(planHandle, fftTotalCount);

err = clfftSetPlanDistance(planHandle, shiftSize, halfedWindowSize);

/* Bake the plan. */

err = clfftBakePlan(planHandle, 1, &queue, NULL, NULL);

}

void calculateFFT()

{

err = clEnqueueWriteBuffer(queue, D_raw, CL_TRUE, 0, rawBufLen * sizeof(POINT_TYPE), H_raw, 0, NULL, NULL);

/* Execute the plan. */

err = clfftEnqueueTransform(planHandle, CLFFT_FORWARD, 1, &queue, 0, NULL, NULL, &D_raw, &D_fft, NULL);

/* Wait for calculations to be finished. */

err = clFinish(queue);

// NO NEED

/* Fetch results of calculations. */

//err = clEnqueueReadBuffer(queue, D_fft, CL_TRUE, 0, bandCount * fftsPerBand * halfedWindowSize * 2 * sizeof(POINT_TYPE), H_sum, 0, NULL, NULL);

}

// call this after you don't need calculateFFT() function anymore. Call before ocl and buffers finalization

void finalize_clFFT()

{

/* Release the plan. */

err = clfftDestroyPlan(&planHandle);

/* Release clFFT library. */

clfftTeardown();

}

// call this before sumVectors() and getMetrics()

void setup_Kernels()

{

kernelSource = getKernelSource();

// Number of work items in each local work group

localSize = 64;

// Number of total work items - localSize must be devisor

globalSize = ceil(resultBandwidth / (float)localSize)*localSize;

// Create the compute program from the source buffer

program = clCreateProgramWithSource(ctx, 1,

(const char **)& kernelSource, NULL, &err);

// Build the program executable

//err = clBuildProgram(program, 0, NULL, NULL, NULL, NULL);

if (clBuildProgram(program, 0, NULL, NULL, NULL, NULL) != CL_SUCCESS)

{

std::cout << "Program Build failed\n";

size_t length;

char buffer[4*2048];

err = clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_LOG, sizeof(buffer), buffer, &length);

std::cout << "--- Build log ---\n " << buffer << std::endl;

exit(1);

}

// Create the compute kernel in the program we wish to run

aggregateKernel = clCreateKernel(program, "aggregateFFTs", &err);

// Set the arguments to our compute kernel

err = clSetKernelArg(aggregateKernel, 0, sizeof(cl_mem), &D_fft);

err |= clSetKernelArg(aggregateKernel, 1, sizeof(cl_mem), &D_sum);

err |= clSetKernelArg(aggregateKernel, 2, sizeof(int), &resultBandwidth);

err |= clSetKernelArg(aggregateKernel, 3, sizeof(int), &totalBandwidth2);

err |= clSetKernelArg(aggregateKernel, 4, sizeof(int), &totalWidth2);

err |= clSetKernelArg(aggregateKernel, 5, sizeof(int), &bandCount);

metricsLocalSize = METRICS_LEN / 2;

metricsGlobalSize = metSmallBatchCount * metBigBatchCount * metricsLocalSize;

// metrics kernel

metricsKernel = clCreateKernel(program, "getMetrics", &err);

// Set the arguments to our metrics kernel

err = clSetKernelArg( metricsKernel, 0, sizeof(cl_mem), &D_raw);

err |= clSetKernelArg(metricsKernel, 1, sizeof(cl_mem), &D_met);

err |= clSetKernelArg(metricsKernel, 2, metPairCount * 2 * sizeof(POINT_TYPE), NULL);

err |= clSetKernelArg(metricsKernel, 3, metPairCount * 2 * sizeof(POINT_TYPE), NULL);

err |= clSetKernelArg(metricsKernel, 4, METRICS_COUNT * sizeof(POINT_TYPE), NULL);

err |= clSetKernelArg(metricsKernel, 5, sizeof(unsigned int), &metPairCount);

err |= clSetKernelArg(metricsKernel, 6, sizeof(unsigned int), &metSmallBatchCount);

err |= clSetKernelArg(metricsKernel, 7, sizeof(unsigned int), &metSmallBatchShift);

err |= clSetKernelArg(metricsKernel, 8, sizeof(unsigned int), &metBigBatchCount);

err |= clSetKernelArg(metricsKernel, 9, sizeof(unsigned int), &metBigBatchShift);

free(kernelSource);

}

void sumVectors()

{

// Execute the kernel over the entire range of the data set

err = clEnqueueNDRangeKernel(queue, aggregateKernel, 1, NULL, &globalSize, &localSize,

0, NULL, NULL);

// Wait for the command queue to get serviced before reading back results

err = clFinish(queue);

// Read the results from the device

err = clEnqueueReadBuffer(queue, D_sum, CL_TRUE, 0, sumBufLen * sizeof(POINT_TYPE), H_sum, 0, NULL, NULL);

}

void getMetrics()

{

// Execute the kernel over the entire range of the data set

err = clEnqueueNDRangeKernel(queue, metricsKernel, 1, NULL, &metricsGlobalSize, &metricsLocalSize,

0, NULL, NULL);

// Wait for the command queue to get serviced before reading back results

err = clFinish(queue);

// Read the results from the device

err = clEnqueueReadBuffer(queue, D_met, CL_TRUE, 0, metBufLen * sizeof(POINT_TYPE), H_met, 0, NULL, NULL);

}

// call this after you don't need sumVectors() and getMetrics() functions anymore. Call before ocl and buffers finalization

void finalize_Kernels()

{

// release OpenCL resources

clReleaseKernel(aggregateKernel);

clReleaseKernel(metricsKernel);

clReleaseProgram(program);

}

}

Размещено на Allbest.ru