Library and event profiling

[TApplencourt]

It's not really an issue; it's just an inconvenience that @colleeneb found when working with oneMKL. I opened this issue to track it so I don't forget, as it's a general inconvenience. We can close if we think it's not the correct place for such discussion.

The problem is that library API returns one event but can use internally multiple commands and hence events. As example, this is a typical case of some mkl functions. Example:

// Assuming in-order for simplicify
auto foo(sycl::queue Q)  { 
 auto e1 =bar(Q)
 auto e2 = bar(Q)
 return e2;
}

 foo.get_profiling_info() # This will not be e1+e2, just e2

This means it's impossible for a user of a library to know how much the library call took on the device.

Maybe one solution will be to add a new event constructor to an event that take of lists of the event to "bundle". Then we can specify get_profiling_info of this event should report the sum of the events used during the construction

Example:

 auto e1 = foo1(Q)
 auto e2 = foo1(Q)
 e3 = sycl::event{ e1, e2 };
 e3. get_profiling_info == (e1.get_profiling_info + e2.get_profiling_info) 

[gmlueck]

The event::get_profiling_info function returns a timestamp, not an elapsed time. Therefore, it's not simply a matter of adding time together from the aggregate events.

If we added an event constructor like you suggest, would you expect info::event_profiling::command_submit to return the submission timestamp of the first constituent event and info::event_profiling::command_end to return the completion timestamp of the last constituent event? This seems like it would be easy enough to implement, but subtracting the two timestamps would not always provide the amount of time that commands were executing on the device. For example, if e1 completed before e2 started, subtracting the timestamps would include the host time in between the two commands.

[TApplencourt]

You are totally correct on both points!

We can either:

• Add a vector of command_{submit,end,start}in the event object. But that is a massive change. Can we create a new aggreateEvent, who is a child of Event and who has vector in place of scalar? Never feels good to add a new object to the spec. Maybe not worth it.
• Add a new "optional" field to provide the elapsed (start-end) time? It's not perfect, but it's what people care most about, IMO.

[AerialMantis]

SYCL WG call:

• Suggestion that this could also be solved via SYCL graphs, idea is to bundle events with profiling information, but the library would still need some way to compose the profiling info.
• Once profiling support in SYCL graphs is more mature, we can look at this
• Still need some way to compose events/profiling info.
• CUDA handles profiling with timestamps but isn't robust as it's at the mercy of scheduling, so time can include switching between streams, can't rely on this unless it's serialized
• Bundling events and profiling information does make this difficult
• Would be useful to allow a barrier on events to compose the profiling info, but this would probably need to be opt-in, avoids introducing a new API

[keryell]

The problem with this as a timing tool is it has an impact on performance. Not only for SYCL: https://github.com/NVIDIA/stdexec/pull/1227 And with chiplets, distributed firmware... it is unlikely to get better in the future.