bgirgis/OpenCL-examples
1
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

added fifth version, which gives very strange results...

Browse Source
This commit is contained in:
Dakota St. Laurent
2015-06-19 09:50:49 -04:00
parent de0533d168
commit 5c5a02b499
1 changed files with 49 additions and 21 deletions
Download Patch File Download Diff File Expand all files Collapse all files

70
example01/main.cpp
View File

@@ -1,6 +1,5 @@
#include <iostream> #include <iostream>
#include <ctime> #include <ctime>
#include <string.h>
#include "../CL/cl.hpp" #include "../CL/cl.hpp"
@@ -42,9 +41,6 @@ double timeAddVectorsCPU(int n, int k) {
int main(int argc, char* argv[]) { int main(int argc, char* argv[]) {
for (int j=0; j<argc; j++)
std::cout << argv[j] << std::endl;
bool verbose; bool verbose;
if (argc == 1 || std::strcmp(argv[1], "0") == 0) if (argc == 1 || std::strcmp(argv[1], "0") == 0)
verbose = true; verbose = true;
@@ -120,23 +116,28 @@ int main(int argc, char* argv[]) {
" v3[i] = v1[i] + v2[i];" " v3[i] = v1[i] + v2[i];"
" }" " }"
"" ""
" void kernel add_single(global const int* v1, global const int* v2, global int* v3) { " " void kernel add_single(global const int* v1, global const int* v2, global int* v3, "
" global const int* constants) { "
" int k = constants[1];"
" int ID = get_global_id(0);" " int ID = get_global_id(0);"
" v3[ID] = v1[ID] + v2[ID];" " for (int i=0; i<k; i++)"
" v3[ID] = v1[ID] + v2[ID];"
" }" " }"
"" // same as add_single, but with the overhead of the looped functions "" // same as add_single, but with the overhead (indexing, determining ratio) of versions 01 and 02
" void kernel add_single_overhead(global const int* v1, global const int* v2, global int* v3," " void kernel add_single_overhead(global const int* v1, global const int* v2, global int* v3,"
" global const int* constants) {" " global const int* constants) {"
" int ID, NUM_GLOBAL_WITEMS, n, ratio, start, stop;" " int ID, NUM_GLOBAL_WITEMS, n, k, ratio, start, stop;"
" ID = get_global_id(0);" " ID = get_global_id(0);"
" NUM_GLOBAL_WITEMS = get_global_size(0);" " NUM_GLOBAL_WITEMS = get_global_size(0);"
" n = constants[0];" " n = constants[0];"
" k = constants[1];"
"" ""
" ratio = (n / NUM_GLOBAL_WITEMS);" " ratio = (n / NUM_GLOBAL_WITEMS);"
" start = ratio * ID;" " start = ratio * ID;"
" stop = ratio * (ID+1);" " stop = ratio * (ID+1);"
"" ""
" v3[ID] = v1[ID] + v2[ID];" " for (int i=0; i<k; i++)"
" v3[ID] = v1[ID] + v2[ID];"
" }"; " }";
sources.push_back({kernel_code.c_str(), kernel_code.length()}); sources.push_back({kernel_code.c_str(), kernel_code.length()});
@@ -163,11 +164,12 @@ int main(int argc, char* argv[]) {
B[i] = n - i - 1; B[i] = n - i - 1;
C[i] = 0; C[i] = 0;
} }
std::clock_t start_time;
// VERSION 1 ========================================== // VERSION 1 ==========================================
// start timer // start timer
double GPUtime1; double GPUtime1;
std::clock_t start_time;
start_time = std::clock(); start_time = std::clock();
// allocate space // allocate space
@@ -192,8 +194,8 @@ int main(int argc, char* argv[]) {
// read result from GPU to here; including for the sake of timing // read result from GPU to here; including for the sake of timing
queue.enqueueReadBuffer(buffer_C, CL_TRUE, 0, sizeof(int)*n, C); queue.enqueueReadBuffer(buffer_C, CL_TRUE, 0, sizeof(int)*n, C);
GPUtime1 = (std::clock() - start_time) / (double) CLOCKS_PER_SEC;
queue.enqueueBarrier(); queue.enqueueBarrier();
GPUtime1 = (std::clock() - start_time) / (double) CLOCKS_PER_SEC;
// VERSION 2 ========================================== // VERSION 2 ==========================================
double GPUtime2; double GPUtime2;
@@ -208,15 +210,15 @@ int main(int argc, char* argv[]) {
queue.enqueueWriteBuffer(buffer_B2, CL_TRUE, 0, sizeof(int)*n, B); queue.enqueueWriteBuffer(buffer_B2, CL_TRUE, 0, sizeof(int)*n, B);
queue.enqueueWriteBuffer(buffer_constants2, CL_TRUE, 0, sizeof(int)*2, constants); queue.enqueueWriteBuffer(buffer_constants2, CL_TRUE, 0, sizeof(int)*2, constants);
add_looped.setArg(0, buffer_A); add_looped.setArg(0, buffer_A2);
add_looped.setArg(1, buffer_B); add_looped.setArg(1, buffer_B2);
add_looped.setArg(2, buffer_C); add_looped.setArg(2, buffer_C2);
add_looped.setArg(3, buffer_constants); add_looped.setArg(3, buffer_constants);
queue.enqueueNDRangeKernel(add, cl::NullRange, cl::NDRange(NUM_GLOBAL_WITEMS), cl::NDRange(32)); queue.enqueueNDRangeKernel(add, cl::NullRange, cl::NDRange(NUM_GLOBAL_WITEMS), cl::NDRange(32));
} }
queue.enqueueReadBuffer(buffer_C2, CL_TRUE, 0, sizeof(int)*n, C); queue.enqueueReadBuffer(buffer_C2, CL_TRUE, 0, sizeof(int)*n, C);
GPUtime2 = (std::clock() - start_time) / (double) CLOCKS_PER_SEC;
queue.enqueueBarrier(); queue.enqueueBarrier();
GPUtime2 = (std::clock() - start_time) / (double) CLOCKS_PER_SEC;
// VERSION 3 ========================================== // VERSION 3 ==========================================
double GPUtime3; double GPUtime3;
@@ -225,16 +227,19 @@ int main(int argc, char* argv[]) {
cl::Buffer buffer_A3(context, CL_MEM_READ_WRITE, sizeof(int)*n); cl::Buffer buffer_A3(context, CL_MEM_READ_WRITE, sizeof(int)*n);
cl::Buffer buffer_B3(context, CL_MEM_READ_WRITE, sizeof(int)*n); cl::Buffer buffer_B3(context, CL_MEM_READ_WRITE, sizeof(int)*n);
cl::Buffer buffer_C3(context, CL_MEM_READ_WRITE, sizeof(int)*n); cl::Buffer buffer_C3(context, CL_MEM_READ_WRITE, sizeof(int)*n);
cl::Buffer buffer_constants3(context, CL_MEM_READ_ONLY, sizeof(int) * 2);
queue.enqueueWriteBuffer(buffer_A3, CL_TRUE, 0, sizeof(int)*n, A); queue.enqueueWriteBuffer(buffer_A3, CL_TRUE, 0, sizeof(int)*n, A);
queue.enqueueWriteBuffer(buffer_B3, CL_TRUE, 0, sizeof(int)*n, B); queue.enqueueWriteBuffer(buffer_B3, CL_TRUE, 0, sizeof(int)*n, B);
queue.enqueueWriteBuffer(buffer_constants3, CL_TRUE, 0, sizeof(int)*2, constants);
add_single.setArg(0, buffer_A3); add_single.setArg(0, buffer_A3);
add_single.setArg(1, buffer_B3); add_single.setArg(1, buffer_B3);
add_single.setArg(2, buffer_C3); add_single.setArg(2, buffer_C3);
add_single.setArg(3, buffer_constants3);
queue.enqueueNDRangeKernel(add_single, cl::NullRange, cl::NDRange(n), cl::NDRange(32)); queue.enqueueNDRangeKernel(add_single, cl::NullRange, cl::NDRange(n), cl::NDRange(32));
// end timer
queue.enqueueReadBuffer(buffer_C3, CL_TRUE, 0, sizeof(int)*n, C); queue.enqueueReadBuffer(buffer_C3, CL_TRUE, 0, sizeof(int)*n, C);
queue.enqueueBarrier();
GPUtime3 = (std::clock() - start_time) / (double) CLOCKS_PER_SEC; GPUtime3 = (std::clock() - start_time) / (double) CLOCKS_PER_SEC;
// VERSION 4 ========================================== // VERSION 4 ==========================================
@@ -255,20 +260,43 @@ int main(int argc, char* argv[]) {
add_single_overhead.setArg(3, buffer_constants4); add_single_overhead.setArg(3, buffer_constants4);
queue.enqueueNDRangeKernel(add_single_overhead, cl::NullRange, cl::NDRange(n), cl::NDRange(32)); queue.enqueueNDRangeKernel(add_single_overhead, cl::NullRange, cl::NDRange(n), cl::NDRange(32));
// end timer
queue.enqueueReadBuffer(buffer_C4, CL_TRUE, 0, sizeof(int)*n, C); queue.enqueueReadBuffer(buffer_C4, CL_TRUE, 0, sizeof(int)*n, C);
queue.enqueueBarrier();
GPUtime4 = (std::clock() - start_time) / (double) CLOCKS_PER_SEC; GPUtime4 = (std::clock() - start_time) / (double) CLOCKS_PER_SEC;
// VERSION 5 ==========================================
double GPUtime5;
start_time = std::clock();
cl::Buffer buffer_A5(context, CL_MEM_READ_WRITE, sizeof(int) * n);
cl::Buffer buffer_B5(context, CL_MEM_READ_WRITE, sizeof(int) * n);
cl::Buffer buffer_C5(context, CL_MEM_READ_WRITE, sizeof(int) * n);
cl::Buffer buffer_constants5(context, CL_MEM_READ_ONLY, sizeof(int) * 2);
queue.enqueueWriteBuffer(buffer_A5, CL_TRUE, 0, sizeof(int)*n, A);
queue.enqueueWriteBuffer(buffer_B5, CL_TRUE, 0, sizeof(int)*n, B);
queue.enqueueWriteBuffer(buffer_constants5, CL_TRUE, 0, sizeof(int)*2, constants);
add_looped.setArg(0, buffer_A5);
add_looped.setArg(1, buffer_B5);
add_looped.setArg(2, buffer_C5);
add_looped.setArg(3, buffer_constants5);
queue.enqueueNDRangeKernel(add_looped, cl::NullRange, cl::NDRange(n), cl::NDRange(32));
queue.enqueueReadBuffer(buffer_C5, CL_TRUE, 0, sizeof(int)*n, C);
queue.enqueueBarrier();
GPUtime5 = (std::clock() - start_time) / (double) CLOCKS_PER_SEC;
// let's compare! // let's compare!
double GPUtimes[4] = {GPUtime1, GPUtime2, GPUtime3, GPUtime4}; const int NUM_VERSIONS = 5;
double GPUtimes[NUM_VERSIONS] = {GPUtime1, GPUtime2, GPUtime3, GPUtime4, GPUtime5};
if (verbose) { if (verbose) {
for (int i=0; i<4; i++) for (int i=0; i<NUM_VERSIONS; i++)
compareResults(CPUtime, GPUtimes[i], i+1); compareResults(CPUtime, GPUtimes[i], i+1);
} else { } else {
std::cout << CPUtime << ","; std::cout << CPUtime << ",";
for (int i=0; i<3; i++) for (int i=0; i<NUM_VERSIONS-1; i++)
std::cout << GPUtimes[i] << ","; std::cout << GPUtimes[i] << ",";
std::cout << GPUtimes[4] << std::endl; std::cout << GPUtimes[NUM_VERSIONS-1] << std::endl;
} }
return 0; return 0;
} }