1 //
2 // Copyright © 2017 Arm Ltd. All rights reserved.
3 // SPDX-License-Identifier: MIT
4 //
5
6 #include "OpenClTimer.hpp"
7
8 #include <armnn/utility/IgnoreUnused.hpp>
9
10 #include <string>
11 #include <sstream>
12
13
14 namespace armnn
15 {
16
OpenClTimer()17 OpenClTimer::OpenClTimer()
18 {
19 }
20
Start()21 void OpenClTimer::Start()
22 {
23 m_Kernels.clear();
24
25 auto interceptor = [this]( cl_command_queue command_queue,
26 cl_kernel kernel,
27 cl_uint work_dim,
28 const size_t *gwo,
29 const size_t *gws,
30 const size_t *lws,
31 cl_uint num_events_in_wait_list,
32 const cl_event * event_wait_list,
33 cl_event * event)
34 {
35 IgnoreUnused(event);
36 cl_int retVal = 0;
37
38 // Get the name of the kernel
39 cl::Kernel retainedKernel(kernel, true);
40 std::stringstream ss;
41 ss << retainedKernel.getInfo<CL_KERNEL_FUNCTION_NAME>();
42
43 // Embed workgroup sizes into the name
44 if(gws != nullptr)
45 {
46 ss << " GWS[" << gws[0] << "," << gws[1] << "," << gws[2] << "]";
47 }
48 if(lws != nullptr)
49 {
50 ss << " LWS[" << lws[0] << "," << lws[1] << "," << lws[2] << "]";
51 }
52
53 cl_event customEvent;
54
55 // Forward to original OpenCl function
56 retVal = m_OriginalEnqueueFunction( command_queue,
57 kernel,
58 work_dim,
59 gwo,
60 gws,
61 lws,
62 num_events_in_wait_list,
63 event_wait_list,
64 &customEvent);
65
66 // Store the Kernel info for later GetMeasurements() call
67 m_Kernels.emplace_back(ss.str(), customEvent);
68
69 return retVal;
70 };
71
72 m_OriginalEnqueueFunction = CLSymbols::get().clEnqueueNDRangeKernel_ptr;
73 CLSymbols::get().clEnqueueNDRangeKernel_ptr = interceptor;
74 }
75
Stop()76 void OpenClTimer::Stop()
77 {
78 CLSymbols::get().clEnqueueNDRangeKernel_ptr = m_OriginalEnqueueFunction;
79 }
80
GetMeasurements() const81 std::vector<Measurement> OpenClTimer::GetMeasurements() const
82 {
83 std::vector<Measurement> measurements;
84
85 cl_command_queue_properties clQueueProperties = CLScheduler::get().queue().getInfo<CL_QUEUE_PROPERTIES>();
86
87 int idx = 0;
88 for (auto& kernel : m_Kernels)
89 {
90 std::string name = std::string(this->GetName()) + "/" + std::to_string(idx++) + ": " + kernel.m_Name;
91
92 double timeUs = 0.0;
93 if((clQueueProperties & CL_QUEUE_PROFILING_ENABLE) != 0)
94 {
95 // Wait for the event to finish before accessing profile results.
96 kernel.m_Event.wait();
97
98 cl_ulong start = kernel.m_Event.getProfilingInfo<CL_PROFILING_COMMAND_START>();
99 cl_ulong end = kernel.m_Event.getProfilingInfo<CL_PROFILING_COMMAND_END>();
100 timeUs = static_cast<double>(end - start) / 1000.0;
101 }
102
103 measurements.emplace_back(name, timeUs, Measurement::Unit::TIME_US);
104 }
105
106 return measurements;
107 }
108
109 } //namespace armnn
110