[Beignet] [PATCH 2/2] runtime: fix some subtle event bugs.

[Yang, Rong R]

The patchset LGTM, thanks.

> -----Original Message-----
> From: Beignet [mailto:beignet-bounces at lists.freedesktop.org] On Behalf Of
> Zhigang Gong
> Sent: Friday, July 11, 2014 10:18 AM
> To: beignet at lists.freedesktop.org
> Cc: Gong, Zhigang
> Subject: [Beignet] [PATCH 2/2] runtime: fix some subtle event bugs.
> 
> This patch fix the following two bugs in event handling.
> 1. When it's time to call a event's user call back function, we need to
>    set the executed to true before the call. As that call back function
>    may call into clReleaseEvent(), and if we don't set the executed status
>    to true, it will enter infinite recursive loop.
> 
> 2. After the user call clEnqueueNDRangeKernel to get a valid event, the
>    user set a call back function to that event, and in that call back
>    function, it will release that event. This scenario is totally correct.
>    But our current event handling doesn't have a deadicated timer thread to
>    update those on-the-fly events' status. Thus those events will not have
>    a chance to get updated, and those call back function will not executed
>    forever. To introduce a complete timer style thread to maintain this type
>    of events is too heavy for this fix release. This patch choose an easy
>    way to work around it. It will make sure the last gpgpu event to be finished
>    before current task to be enqueued.
> 
> After this patch, most of the OpenCV 3.0 cases could run smoothly without any
> serious issue.
> 
> Signed-off-by: Zhigang Gong <zhigang.gong at intel.com>
> ---
>  src/cl_api.c           |  2 +-
>  src/cl_command_queue.c | 11 +++++++++++
>  src/cl_event.c         | 12 ++++++------
>  src/cl_event.h         |  2 +-
>  4 files changed, 19 insertions(+), 8 deletions(-)
> 
> diff --git a/src/cl_api.c b/src/cl_api.c index 8759027..177a7e8 100644
> --- a/src/cl_api.c
> +++ b/src/cl_api.c
> @@ -1365,7 +1365,7 @@ clGetEventInfo(cl_event      event,
>    } else if (param_name == CL_EVENT_COMMAND_TYPE) {
>      FILL_GETINFO_RET (cl_command_type, 1, &event->type, CL_SUCCESS);
>    } else if (param_name == CL_EVENT_COMMAND_EXECUTION_STATUS) {
> -    cl_event_update_status(event);
> +    cl_event_update_status(event, 0);
>      FILL_GETINFO_RET (cl_int, 1, &event->status, CL_SUCCESS);
>    } else if (param_name == CL_EVENT_REFERENCE_COUNT) {
>      cl_uint ref = event->ref_n;
> diff --git a/src/cl_command_queue.c b/src/cl_command_queue.c index
> d9718bf..d45e92f 100644
> --- a/src/cl_command_queue.c
> +++ b/src/cl_command_queue.c
> @@ -75,6 +75,10 @@ cl_command_queue_delete(cl_command_queue queue)
>    assert(queue);
>    if (atomic_dec(&queue->ref_n) != 1) return;
> 
> +  // If there is a valid last event, we need to give it a chance to  //
> + call the call-back function.
> +  if (queue->last_event && queue->last_event->user_cb)
> +    cl_event_update_status(queue->last_event, 1);
>    /* Remove it from the list */
>    assert(queue->ctx);
>    pthread_mutex_lock(&queue->ctx->queue_lock);
> @@ -454,6 +458,13 @@ cl_command_queue_flush(cl_command_queue
> queue)  {
>    GET_QUEUE_THREAD_GPGPU(queue);
>    cl_command_queue_flush_gpgpu(queue, gpgpu);
> +  // As we don't have a deadicate timer thread to take care the
> + possible  // event which has a call back function registerred and the
> + event will  // be released at the call back function, no other
> + function will access  // the event any more. If we don't do this here,
> + we will leak that event  // and all the corresponding buffers which is really
> bad.
> +  if (queue->last_event && queue->last_event->user_cb)
> +    cl_event_update_status(queue->last_event, 1);
>    if (queue->current_event)
>      cl_event_flush(queue->current_event);
>    cl_invalid_thread_gpgpu(queue);
> diff --git a/src/cl_event.c b/src/cl_event.c index d40881a..99e60eb 100644
> --- a/src/cl_event.c
> +++ b/src/cl_event.c
> @@ -55,6 +55,7 @@ void cl_event_flush(cl_event event)
>      event->gpgpu = NULL;
>    }
>    cl_gpgpu_event_flush(event->gpgpu_event);
> +  event->queue->last_event = event;
>  }
> 
>  cl_event cl_event_new(cl_context ctx, cl_command_queue queue,
> cl_command_type type, cl_bool emplict) @@ -95,8 +96,6 @@ cl_event
> cl_event_new(cl_context ctx, cl_command_queue queue, cl_command_type ty
>    event->enqueue_cb = NULL;
>    event->waits_head = NULL;
>    event->emplict = emplict;
> -  if(queue && event->gpgpu_event)
> -    queue->last_event = event;
> 
>  exit:
>    return event;
> @@ -111,7 +110,7 @@ void cl_event_delete(cl_event event)
>    if (UNLIKELY(event == NULL))
>      return;
> 
> -  cl_event_update_status(event);
> +  cl_event_update_status(event, 0);
> 
>    if (atomic_dec(&event->ref_n) > 1)
>      return;
> @@ -124,6 +123,7 @@ void cl_event_delete(cl_event event)
>    while(event->user_cb) {
>      cb = event->user_cb;
>      if(cb->executed == CL_FALSE) {
> +      cb->executed = CL_TRUE;
>        cb->pfn_notify(event, event->status, cb->user_data);
>      }
>      event->user_cb = cb->next;
> @@ -443,8 +443,8 @@ void cl_event_set_status(cl_event event, cl_int status)
>    user_cb = event->user_cb;
>    while(user_cb) {
>      if(user_cb->status >= status) {
> -      user_cb->pfn_notify(event, event->status, user_cb->user_data);
>        user_cb->executed = CL_TRUE;
> +      user_cb->pfn_notify(event, event->status, user_cb->user_data);
>      }
>      user_cb = user_cb->next;
>    }
> @@ -492,12 +492,12 @@ void cl_event_set_status(cl_event event, cl_int
> status)
>    event->waits_head = NULL;
>  }
> 
> -void cl_event_update_status(cl_event event)
> +void cl_event_update_status(cl_event event, int wait)
>  {
>    if(event->status <= CL_COMPLETE)
>      return;
>    if((event->gpgpu_event) &&
> -     (cl_gpgpu_event_update_status(event->gpgpu_event, 0) ==
> command_complete))
> +     (cl_gpgpu_event_update_status(event->gpgpu_event, wait) ==
> + command_complete))
>      cl_event_set_status(event, CL_COMPLETE);  }
> 
> diff --git a/src/cl_event.h b/src/cl_event.h index 3c23d74..cfe5ddd 100644
> --- a/src/cl_event.h
> +++ b/src/cl_event.h
> @@ -89,7 +89,7 @@ void cl_event_new_enqueue_callback(cl_event,
> enqueue_data *, cl_uint, const cl_e
>  /* Set the event status and call all callbacks */  void
> cl_event_set_status(cl_event, cl_int);
>  /* Check and update event status */
> -void cl_event_update_status(cl_event);
> +void cl_event_update_status(cl_event, cl_int);
>  /* Create the marker event */
>  cl_int cl_event_marker_with_wait_list(cl_command_queue, cl_uint, const
> cl_event *,  cl_event*);
>  /* Create the barrier event */
> --
> 1.8.3.2
> 
> _______________________________________________
> Beignet mailing list
> Beignet at lists.freedesktop.org
> http://lists.freedesktop.org/mailman/listinfo/beignet