OpenCL device timer

thysol

thysol

BIOS-Overclocker(in)
Hi,
Ich moechte gerne die Zeit messen die ein Device braucht um einen Kernel auszufuehren. Das mache ich so:

ret = clGetEventProfilingInfo(event,CL_PROFILING_COMMAND_START,sizeof(long long),&kernelsStartTime,NULL);
ret = clGetEventProfilingInfo(event,CL_PROFILING_COMMAND_END,sizeof(long long),&kernelsEndTime,NULL);
Zum Vergrößern anklicken....
Allerdings ist die Zeit die gemessen wird immer gleich, auch wenn der Kernel gefuehlt mehrere Sekunden zum ausfuehren braucht gibt der Timer immer 0.00813 Sekunden als Zeit an. Diese Zeit kriege ich immer, egal ob ich den Kernel auf der CPU oder GPU ausfuehre. Das kann ja wohl nicht sein. Was habe ich falsch gemacht?

Hier ist der gesamte Code:

#include <stdio.h>
#include <stdlib.h>
#include <pthread.h>
#ifdef __APPLE__
#include <OpenCL/opencl.h>
#else
#include <CL/cl.h>
#endif
#define MAX_SOURCE_SIZE (0x100000)
void *threadFunc(void *arg)
{
cl_double totalKernelTime;
long long kernelsStartTime;
long long kernelsEndTime;
cl_event event;
int i;
const int LIST_SIZE = 1000000;
float *A = (float*)malloc(sizeof(float)*LIST_SIZE);
float *B = (float*)malloc(sizeof(float)*LIST_SIZE);
float *C = (float*)malloc(sizeof(float)*LIST_SIZE);
float *D = (float*)malloc(sizeof(float)*LIST_SIZE);
float *E = (float*)malloc(sizeof(float)*LIST_SIZE);
float *F = (float*)malloc(sizeof(float)*LIST_SIZE);
float *G = (float*)malloc(sizeof(float)*LIST_SIZE);
float *H = (float*)malloc(sizeof(float)*LIST_SIZE);
for(i = 0; i < LIST_SIZE; i++) {
A = 1000 + (i * 0.099);
B = 200000 + (i * 9.8);
C = 6378140 + B;
}
FILE *fp;
char *source_str;
size_t source_size;
fp = fopen("vector_add_kernel2.cl", "r");
if (!fp) {
fprintf(stderr, "Failed to load kernel.\n");
exit(1);
}
source_str = (char*)malloc(MAX_SOURCE_SIZE);
source_size = fread( source_str, 1, MAX_SOURCE_SIZE, fp);
fclose( fp );
cl_platform_id platform_id = NULL;
cl_device_id device_id = NULL;
cl_uint ret_num_devices;
cl_uint ret_num_platforms;
cl_int ret = clGetPlatformIDs(1, &platform_id, &ret_num_platforms);
ret = clGetDeviceIDs( platform_id, CL_DEVICE_TYPE_GPU, 1,
&device_id, &ret_num_devices);
cl_context context = clCreateContext( NULL, 1, &device_id, NULL, NULL, &ret);
cl_command_queue command_queue = clCreateCommandQueue(context, device_id, CL_QUEUE_PROFILING_ENABLE, &ret);
cl_mem a_mem_obj = clCreateBuffer(context, CL_MEM_READ_ONLY,
LIST_SIZE * sizeof(float), NULL, &ret);
cl_mem b_mem_obj = clCreateBuffer(context, CL_MEM_READ_ONLY,
LIST_SIZE * sizeof(float), NULL, &ret);
cl_mem c_mem_obj = clCreateBuffer(context, CL_MEM_READ_ONLY,
LIST_SIZE * sizeof(float), NULL, &ret);
cl_mem d_mem_obj = clCreateBuffer(context, CL_MEM_WRITE_ONLY,
LIST_SIZE * sizeof(float), NULL, &ret);
cl_mem e_mem_obj = clCreateBuffer(context, CL_MEM_WRITE_ONLY,
LIST_SIZE * sizeof(float), NULL, &ret);
cl_mem f_mem_obj = clCreateBuffer(context, CL_MEM_WRITE_ONLY,
LIST_SIZE * sizeof(float), NULL, &ret);
cl_mem g_mem_obj = clCreateBuffer(context, CL_MEM_WRITE_ONLY,
LIST_SIZE * sizeof(float), NULL, &ret);
cl_mem h_mem_obj = clCreateBuffer(context, CL_MEM_WRITE_ONLY,
LIST_SIZE * sizeof(float), NULL, &ret);
ret = clEnqueueWriteBuffer(command_queue, a_mem_obj, CL_TRUE, 0,
LIST_SIZE * sizeof(float), A, 0, NULL, NULL);
ret = clEnqueueWriteBuffer(command_queue, b_mem_obj, CL_TRUE, 0,
LIST_SIZE * sizeof(float), B, 0, NULL, NULL);
ret = clEnqueueWriteBuffer(command_queue, c_mem_obj, CL_TRUE, 0,
LIST_SIZE * sizeof(float), C, 0, NULL, NULL);
cl_program program = clCreateProgramWithSource(context, 1,
(const char **)&source_str, (const size_t *)&source_size, &ret);
ret = clBuildProgram(program, 1, &device_id, NULL, NULL, NULL);
cl_kernel kernel = clCreateKernel(program, "vector_add", &ret);
ret = clSetKernelArg(kernel, 0, sizeof(cl_mem), (void *)&a_mem_obj);
ret = clSetKernelArg(kernel, 1, sizeof(cl_mem), (void *)&b_mem_obj);
ret = clSetKernelArg(kernel, 2, sizeof(cl_mem), (void *)&c_mem_obj);
ret = clSetKernelArg(kernel, 3, sizeof(cl_mem), (void *)&d_mem_obj);
ret = clSetKernelArg(kernel, 4, sizeof(cl_mem), (void *)&e_mem_obj);
ret = clSetKernelArg(kernel, 5, sizeof(cl_mem), (void *)&f_mem_obj);
ret = clSetKernelArg(kernel, 6, sizeof(cl_mem), (void *)&g_mem_obj);
ret = clSetKernelArg(kernel, 7, sizeof(cl_mem), (void *)&h_mem_obj);
size_t global_item_size = LIST_SIZE; // Process the entire lists
size_t local_item_size = 50; // Process one item at a time
ret = clEnqueueNDRangeKernel(command_queue, kernel, 1, NULL,
&global_item_size, &local_item_size, 0, NULL, &event);
ret = clWaitForEvents(1, &event);
ret = clEnqueueReadBuffer(command_queue, d_mem_obj, CL_TRUE, 0,
LIST_SIZE * sizeof(float), D, 0, NULL, NULL);
ret = clEnqueueReadBuffer(command_queue, e_mem_obj, CL_TRUE, 0,
LIST_SIZE * sizeof(float), E, 0, NULL, NULL);
ret = clEnqueueReadBuffer(command_queue, f_mem_obj, CL_TRUE, 0,
LIST_SIZE * sizeof(float), F, 0, NULL, NULL);
ret = clEnqueueReadBuffer(command_queue, g_mem_obj, CL_TRUE, 0,
LIST_SIZE * sizeof(float), G, 0, NULL, NULL);
ret = clEnqueueReadBuffer(command_queue, h_mem_obj, CL_TRUE, 0,
LIST_SIZE * sizeof(float), H, 0, NULL, NULL);
ret = clGetEventProfilingInfo(event,CL_PROFILING_COMMAND_START,sizeof(long long),&kernelsStartTime,NULL);
ret = clGetEventProfilingInfo(event,CL_PROFILING_COMMAND_END,sizeof(long long),&kernelsEndTime,NULL);
totalKernelTime = (double)(kernelsEndTime - kernelsStartTime)/1e9;
clReleaseEvent(event);
printf("%f\n", totalKernelTime);
ret = clFlush(command_queue);
ret = clFinish(command_queue);
ret = clReleaseKernel(kernel);
ret = clReleaseProgram(program);
ret = clReleaseMemObject(a_mem_obj);
ret = clReleaseMemObject(b_mem_obj);
ret = clReleaseMemObject(c_mem_obj);
ret = clReleaseMemObject(d_mem_obj);
ret = clReleaseMemObject(e_mem_obj);
ret = clReleaseMemObject(f_mem_obj);
ret = clReleaseMemObject(g_mem_obj);
ret = clReleaseMemObject(h_mem_obj);
ret = clReleaseCommandQueue(command_queue);
ret = clReleaseContext(context);
free(A);
free(B);
free(C);
free(D);
free(E);
free(F);
free(G);
free(H);
return 0;
}
int main(void) {
pthread_t pth; // this is our thread identifier
pthread_create(&pth,NULL,threadFunc,"processing...");
cl_double totalKernelTime;
long long kernelsStartTime;
long long kernelsEndTime;
cl_event event;
int i;
const int LIST_SIZE = 1000000;
float *A = (float*)malloc(sizeof(float)*LIST_SIZE);
float *B = (float*)malloc(sizeof(float)*LIST_SIZE);
float *C = (float*)malloc(sizeof(float)*LIST_SIZE);
float *D = (float*)malloc(sizeof(float)*LIST_SIZE);
float *E = (float*)malloc(sizeof(float)*LIST_SIZE);
float *F = (float*)malloc(sizeof(float)*LIST_SIZE);
float *G = (float*)malloc(sizeof(float)*LIST_SIZE);
float *H = (float*)malloc(sizeof(float)*LIST_SIZE);
for(i = 0; i < LIST_SIZE; i++) {
A = 1000 + (i * 0.099);
B = 200000 + (i * 9.8);
C = 6378140 + B;
}
FILE *fp;
char *source_str;
size_t source_size;
fp = fopen("vector_add_kernel2.cl", "r");
if (!fp) {
fprintf(stderr, "Failed to load kernel.\n");
exit(1);
}
source_str = (char*)malloc(MAX_SOURCE_SIZE);
source_size = fread( source_str, 1, MAX_SOURCE_SIZE, fp);
fclose( fp );
cl_platform_id platform_id = NULL;
cl_device_id device_id = NULL;
cl_uint ret_num_devices;
cl_uint ret_num_platforms;
cl_int ret = clGetPlatformIDs(1, &platform_id, &ret_num_platforms);
ret = clGetDeviceIDs( platform_id, CL_DEVICE_TYPE_CPU, 1,
&device_id, &ret_num_devices);
cl_context context = clCreateContext( NULL, 1, &device_id, NULL, NULL, &ret);
cl_command_queue command_queue = clCreateCommandQueue(context, device_id, 0, &ret);
cl_mem a_mem_obj = clCreateBuffer(context, CL_MEM_READ_ONLY,
LIST_SIZE * sizeof(float), NULL, &ret);
cl_mem b_mem_obj = clCreateBuffer(context, CL_MEM_READ_ONLY,
LIST_SIZE * sizeof(float), NULL, &ret);
cl_mem c_mem_obj = clCreateBuffer(context, CL_MEM_READ_ONLY,
LIST_SIZE * sizeof(float), NULL, &ret);
cl_mem d_mem_obj = clCreateBuffer(context, CL_MEM_WRITE_ONLY,
LIST_SIZE * sizeof(float), NULL, &ret);
cl_mem e_mem_obj = clCreateBuffer(context, CL_MEM_WRITE_ONLY,
LIST_SIZE * sizeof(float), NULL, &ret);
cl_mem f_mem_obj = clCreateBuffer(context, CL_MEM_WRITE_ONLY,
LIST_SIZE * sizeof(float), NULL, &ret);
cl_mem g_mem_obj = clCreateBuffer(context, CL_MEM_WRITE_ONLY,
LIST_SIZE * sizeof(float), NULL, &ret);
cl_mem h_mem_obj = clCreateBuffer(context, CL_MEM_WRITE_ONLY,
LIST_SIZE * sizeof(float), NULL, &ret);
ret = clEnqueueWriteBuffer(command_queue, a_mem_obj, CL_TRUE, 0,
LIST_SIZE * sizeof(float), A, 0, NULL, NULL);
ret = clEnqueueWriteBuffer(command_queue, b_mem_obj, CL_TRUE, 0,
LIST_SIZE * sizeof(float), B, 0, NULL, NULL);
ret = clEnqueueWriteBuffer(command_queue, c_mem_obj, CL_TRUE, 0,
LIST_SIZE * sizeof(float), C, 0, NULL, NULL);
cl_program program = clCreateProgramWithSource(context, 1,
(const char **)&source_str, (const size_t *)&source_size, &ret);
ret = clBuildProgram(program, 1, &device_id, NULL, NULL, NULL);
cl_kernel kernel = clCreateKernel(program, "vector_add", &ret);
ret = clSetKernelArg(kernel, 0, sizeof(cl_mem), (void *)&a_mem_obj);
ret = clSetKernelArg(kernel, 1, sizeof(cl_mem), (void *)&b_mem_obj);
ret = clSetKernelArg(kernel, 2, sizeof(cl_mem), (void *)&c_mem_obj);
ret = clSetKernelArg(kernel, 3, sizeof(cl_mem), (void *)&d_mem_obj);
ret = clSetKernelArg(kernel, 4, sizeof(cl_mem), (void *)&e_mem_obj);
ret = clSetKernelArg(kernel, 5, sizeof(cl_mem), (void *)&f_mem_obj);
ret = clSetKernelArg(kernel, 6, sizeof(cl_mem), (void *)&g_mem_obj);
ret = clSetKernelArg(kernel, 7, sizeof(cl_mem), (void *)&h_mem_obj);
size_t global_item_size = LIST_SIZE; // Process the entire lists
size_t local_item_size = 50; // Process one item at a time
ret = clEnqueueNDRangeKernel(command_queue, kernel, 1, NULL,
&global_item_size, &local_item_size, 0, NULL, &event);
ret = clWaitForEvents(1, &event);
ret = clEnqueueReadBuffer(command_queue, d_mem_obj, CL_TRUE, 0,
LIST_SIZE * sizeof(float), D, 0, NULL, NULL);
ret = clEnqueueReadBuffer(command_queue, e_mem_obj, CL_TRUE, 0,
LIST_SIZE * sizeof(float), E, 0, NULL, NULL);
ret = clEnqueueReadBuffer(command_queue, f_mem_obj, CL_TRUE, 0,
LIST_SIZE * sizeof(float), F, 0, NULL, NULL);
ret = clEnqueueReadBuffer(command_queue, g_mem_obj, CL_TRUE, 0,
LIST_SIZE * sizeof(float), G, 0, NULL, NULL);
ret = clEnqueueReadBuffer(command_queue, h_mem_obj, CL_TRUE, 0,
LIST_SIZE * sizeof(float), H, 0, NULL, NULL);
pthread_join(pth, NULL /* void ** return value could go here */);
ret = clGetEventProfilingInfo(event,CL_PROFILING_COMMAND_START,sizeof(long long),&kernelsStartTime,NULL);
ret = clGetEventProfilingInfo(event,CL_PROFILING_COMMAND_END,sizeof(long long),&kernelsEndTime,NULL);
totalKernelTime = (double)(kernelsEndTime - kernelsStartTime)/1e9;
clReleaseEvent(event);
printf("%f\n", totalKernelTime);
ret = clFlush(command_queue);
ret = clFinish(command_queue);
ret = clReleaseKernel(kernel);
ret = clReleaseProgram(program);
ret = clReleaseMemObject(a_mem_obj);
ret = clReleaseMemObject(b_mem_obj);
ret = clReleaseMemObject(c_mem_obj);
ret = clReleaseMemObject(d_mem_obj);
ret = clReleaseMemObject(e_mem_obj);
ret = clReleaseMemObject(f_mem_obj);
ret = clReleaseMemObject(g_mem_obj);
ret = clReleaseMemObject(h_mem_obj);
ret = clReleaseCommandQueue(command_queue);
ret = clReleaseContext(context);
free(A);
free(B);
free(C);
free(D);
free(E);
free(F);
free(G);
free(H);
return 0;
}
Zum Vergrößern anklicken....
 
Zuletzt bearbeitet:
Thysol, du misst da wenn ich es richtig in Erinnerung habe nur die Zeit, die benötigt wird, um den Kernel zu starten.

Erstell ein Event, und mess, wann die Queue wieder leer ist. queue.finished() war glaub ich eine Möglichkeit.

Der Knackpunkt sollte eben sein, das die Sache eben nicht blocking ausgeführt wird.

Das hier funktioniert aber auf jeden Fall:

// Create command queue using the first device
! ! CommandQueue queue = CommandQueue( context, devices[0], CL_QUEUE_PROFILING_ENABLE );
// Run kernel
! ! Event event;
! ! queue.enqueueNDRangeKernel(! kernel,! NullRange, global, local, NULL, &event );
! ! event.wait();
! !
! ! cl_ulong queued, submit, start, end;
! ! event.getProfilingInfo( CL_PROFILING_COMMAND_QUEUED, &queued );!
! ! event.getProfilingInfo( CL_PROFILING_COMMAND_SUBMIT, &submit );!
! ! event.getProfilingInfo( CL_PROFILING_COMMAND_START, &start );!
! ! event.getProfilingInfo( CL_PROFILING_COMMAND_END, &end );!
! ! cout << "QUEUED: " << queued << endl
! ! << !"SUBMIT: " << submit << endl
! ! <<!"START: " << start << endl
! ! <<!"END: " << end << endl
! ! <<!"runtime: " << (end - start) *1.e-9 << " secs" << endl;
Zum Vergrößern anklicken....
 
Danke fuer die Hilfe. Jetzt kann ich die Kernel ausfuehrungszeit genau messen. Allerdings habe ich jetzt noch ein Problem. Ich moechte 2 Kernel in parallel ausfuehren, also einen auf der CPU und einen auf der GPU. Daher habe ich in C noch einen Thread erstellt wo dann nochmal das OpenCL Zeug gestartet wird nur auf dem anderen Device. Allerdings klappt das nicht, es wird zwar kein Fehler zurueckgeworfen aber immer wenn ich beide Kernel parallel ausfuehre wirft mir OpenCL voellig falsche Zeitwerte die ueberhaupt nicht stimmen koennen, und der Kernel wird auch gefuehlt viel zu schnell ausgefuehrt, daher gehe ich davon aus das er aus irgendeinem Grund ueberhaupt nicht ausgefuehrt wird. Der andere Kernel wird aber immer perfekt ausgefuehrt, die Zeitwerte stimmen auch. Was noch merkwuerdiger ist, es ist immer unterschiedlich welcher Kernel korrekt ausgefuehrt wird. Mal laeuft der auf der CPU korrekt, mal der auf der GPU. Wenn ich den C multi-threading kram weglasse klappt immer alles perfekt, nur wenn ich die Kernel in parallel ausfuehren moechte gibts Probleme. Was mache ich falsch?
 
OpenCL ist nicht Thread-safe. Daher klappt das nicht ;)

Also zumindest soweit ich mich erinnern kann ist es nicht Thread-safe.

Versuch es mal mit einer eigenen queue für CPU und GPU. Dann sollte das klappen. Zumindest wenn du nur einen Thread nutzt. Bei mehreren musst du mal schauen.
 
Skysnake schrieb:
OpenCL ist nicht Thread-safe. Daher klappt das nicht ;)

Also zumindest soweit ich mich erinnern kann ist es nicht Thread-safe.

Versuch es mal mit einer eigenen queue für CPU und GPU. Dann sollte das klappen. Zumindest wenn du nur einen Thread nutzt. Bei mehreren musst du mal schauen.
Zum Vergrößern anklicken....

Wenn ich die Kernels so ausfuehre:

for(unsigned int i = 0; i < ciDeviceCount; i++)
{
ciErrNum = clEnqueueNDRangeKernel(commandQueue, reduceKernel, 1, 0, globalWorkSize, localWorkSize,
0, NULL, &GPUExecution);
oclCheckError(ciErrNum, CL_SUCCESS);
}
Zum Vergrößern anklicken....


Wuerden die dann nicht einem nach dem anderen ausgefuehrt? Oder wuerde das parallel auf der CPU und GPU laufen?
 
das sollte dann parallel ausgeführt werden. Zumindest soweit, die CPU die Tasks in die Queue stopfen kann. Braucht ja auch einiges an Zeit.
 
ja geil, und was macht man da jetzt? :ugly:


ganz im ernst, erzähl mir bitte die Lösung für das Problem :Daumen:
 
ja wäre wirklich nett. In 2 Wochen kann ich auch mal meinen Dozenten fagen, was da Sache ist. Der macht genau das eigentlich ständig :ugly:
 
Sodele, ich steh gerade auch wieder vor dem Problem, dass ich die Ausführungszeit eines Kernels messen muss.

Und TATA ich flieg voll auf die Schnauze weils nicht funktioniert :wall:

Also ich nutze events, und da dann die Times, wann ich das in die Queue rein steckt, und wann das Ergebnis raus kommt.

Ich verwende auch ein clFinish, hilft aber alles nichts. Ich bekomme immer NUR! den gleichen Wert raus, der dann bei 10^13 irgendwo landet, was natürlich totaler quatsch ist, vor allem weil immer genau die gleichen Werte zurückgegeben werden -.-

Das "lustige" an der Sache ist folgendes: Wenn ich in VS2010 das Ding normal im Debugging oder Release mode ausführe kommt ja der Quark raus, wenn ich aber das plugin für VS2010 verwende, mit dem automatisch von allem die profiles angelegt werden, dann werden bei mir auch die richtigen Werte ausgegeben :ugly:

DAS ist doch mal geil oder? Jemand ne Idee, was ich falsch mach?

EDIT:

So als ich das hier geschrieben habe, ist mir natürlich gleich die Lösung gekommen. Ich hab eine Zeile übersehen :klatsch:

Code: 
command_queue[i]=clCreateCommandQueue(context[i], device[i], CL_QUEUE_PROFILING_ENABLE , &err);

Man muss ja das Profiling erst aktivieren beim erstellen der Queue :wall:

Sowas vergisst man wirklich VIEL zu einfach. Funktioniert aber ohne Probleme jetzt :daumen:
 
Einloggen o. Registrieren zum Antworten.

Ähnliche Themen

H
Neuer PC-Aufbau - und einige Spiele starten nicht mehr!
2
Antworten
33
Aufrufe
5K
Hauruck
H
H
Mircolags / DPS Latency Moon erklärung.
Antworten
0
Aufrufe
1K
herba
H
CosmoCortney
<Qt> Wie kann ich eine Assembler-Datei einbinden und kompilieren?
Antworten
0
Aufrufe
777
CosmoCortney
CosmoCortney
C
AMD Rayzen 7 1700x
Antworten
3
Aufrufe
1K
chira1337
C
D
System Service Exception
2
Antworten
22
Aufrufe
6K
Daryl_Dixon
D
Oben Unten
Zurück