The cl_event_new function in cl_event.c returns an event with an initial ref_n value of 2, which causes it to never be deleted, when used correctly. The problem is that upon initialization, the ref_n field gets a value of 1, but in the end the "constructor" will call the cl_event_add_ref on the event, which will increase the value of ref_n to 2, and this is obviously not correct. So either the value to which it is initialized in the beginning should be 0, or the cl_event_add_ref function should not be called on it in cl_event_new. The bug can be demonstrated with this little code for example: https://gist.github.com/torokati44/ca6da2112c0d784f4e79 It does nothing, but executes an empty kernel a whole lot of times in succession, and it seems like there is nothing wrong with it. After around 65000 (which is a value that made me think of an overflow) executions, the result of clEnqueueNDRangeKernel will be CL_OUT_OF_RESOURCES, but only if the context is created with profiling enabled. This is because the timestamp buffer of the internal gpgpu event (which is allocated only if profiling is enabled) will not be freed for any of the created events. The event structures themselves leak too, but those are not so noticeable. Applying either one of the two fixes suggested above makes the problem disappear. So yeah, please fix it, I had a wonderful 3 days hunting it down. :) Attila
Well, further testing has shown that simply applying one of my suggestions causes some other, quite strange errors in cases that are more complicated than the linked little example. (Like a segmentation fault upon entering the clWaitForEvents function, but before executing its first line.) Also, some utests fail. I'm not sure if it is just a local build error, but you will surely know better. And the leakage should still be taken care of somehow. Thank you in advance.
(In reply to TÖRÖK Attila from comment #0) > The cl_event_new function in cl_event.c returns an event with an initial > ref_n value of 2, which causes it to never be deleted, when used correctly. > > The problem is that upon initialization, the ref_n field gets a value of 1, > but in the end the "constructor" will call the cl_event_add_ref on the > event, which will increase the value of ref_n to 2, and this is obviously > not correct. > > So either the value to which it is initialized in the beginning should be 0, > or the cl_event_add_ref function should not be called on it in cl_event_new. > > The bug can be demonstrated with this little code for example: > https://gist.github.com/torokati44/ca6da2112c0d784f4e79 > It does nothing, but executes an empty kernel a whole lot of times in > succession, and it seems like there is nothing wrong with it. > > After around 65000 (which is a value that made me think of an overflow) > executions, the result of clEnqueueNDRangeKernel will be > CL_OUT_OF_RESOURCES, but only if the context is created with profiling > enabled. > > This is because the timestamp buffer of the internal gpgpu event (which is > allocated only if profiling is enabled) will not be freed for any of the > created events. The event structures themselves leak too, but those are not > so noticeable. > > Applying either one of the two fixes suggested above makes the problem > disappear. > > So yeah, please fix it, I had a wonderful 3 days hunting it down. :) > > Attila This is not a bug, you need to release the event after the clWaitForEvents.
Oh, yes. You are right. How embarrassing. Sorry for wasting your time. I guess I'll have to RTFM next time.
Use of freedesktop.org services, including Bugzilla, is subject to our Code of Conduct. How we collect and use information is described in our Privacy Policy.