mirror/tracy
1
0
Fork 0
mirror of https://github.com/wolfpld/tracy.git synced 2024-11-22 14:44:34 +00:00
Code Issues Packages Projects Releases Wiki Activity

add log and simple statistics for async kernel

Browse Source
This commit is contained in:
Cloud Han 2022-01-18 13:51:27 +08:00
parent f6894c8d6c
commit e28c562a20
1 changed files with 25 additions and 1 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

26
examples/OpenCLVectorAdd/OpenCLVectorAdd.cpp
View File

@ -1,3 +1,4 @@
#include <algorithm>
#include <iostream> #include <iostream>
#include <cassert> #include <cassert>
#include <string> #include <string>
@ -137,7 +138,10 @@ int main()
} }
} }
for (int i = 0; i < 100; ++i) const int numLaunchs = 100;
std::vector<cl_event> kernelLaunchEvts;
kernelLaunchEvts.reserve(numLaunchs);
for (int i = 0; i < numLaunchs; ++i)
{ {
int n_value = static_cast<int>(N); int n_value = static_cast<int>(N);
ZoneScopedN("VectorAdd Kernel Launch"); ZoneScopedN("VectorAdd Kernel Launch");
@ -151,6 +155,8 @@ int main()
cl_event vectorAddKernelEvent; cl_event vectorAddKernelEvent;
CL_ASSERT(clEnqueueNDRangeKernel(commandQueue, vectorAddKernel, 1, nullptr, &N, nullptr, 0, nullptr, &vectorAddKernelEvent)); CL_ASSERT(clEnqueueNDRangeKernel(commandQueue, vectorAddKernel, 1, nullptr, &N, nullptr, 0, nullptr, &vectorAddKernelEvent));
TracyCLZoneSetEvent(vectorAddKernelEvent); TracyCLZoneSetEvent(vectorAddKernelEvent);
CL_ASSERT(clRetainEvent(vectorAddKernelEvent));
kernelLaunchEvts.push_back(vectorAddKernelEvent);
std::cout << "VectorAdd Kernel Enqueued" << std::endl; std::cout << "VectorAdd Kernel Enqueued" << std::endl;
} }
@ -164,6 +170,24 @@ int main()
} }
CL_ASSERT(clFinish(commandQueue)); CL_ASSERT(clFinish(commandQueue));
std::vector<float> durations(kernelLaunchEvts.size());
for (int i=0; i<kernelLaunchEvts.size(); i++) {
cl_event evt = kernelLaunchEvts[i];
cl_ulong start;
cl_ulong end;
CL_ASSERT(clGetEventProfilingInfo(evt, CL_PROFILING_COMMAND_START, sizeof(cl_ulong), &start, nullptr));
CL_ASSERT(clGetEventProfilingInfo(evt, CL_PROFILING_COMMAND_END, sizeof(cl_ulong), &end, nullptr));
CL_ASSERT(clReleaseEvent(evt));
durations[i] = (end - start) * 0.001;
std::cout << "VectorAdd Kernel " << i << " tooks " << static_cast<int>(durations[i]) << "us" << std::endl;
};
float avg = std::accumulate(durations.cbegin(), durations.cend(), 0.0f) / durations.size();
float stddev2 = std::accumulate(durations.cbegin(), durations.cend(), 0.0f, [avg](const float& acc, const float& v) {
auto d = v - avg;
return acc + d*d;
}) / (durations.size() - 1.0f);
std::cout << "VectorAdd runtime avg: " << avg << "us, std: " << sqrt(stddev2) << "us over " << numLaunchs << " runs." << std::endl;
TracyCLCollect(tracyCLCtx); TracyCLCollect(tracyCLCtx);