[Mesa-dev] [PATCH v2 2/4] clover: Pass work_dim parameter of clEnqueueNDRangeKernel() to driver

[Jan Vesely]

On Wed, 2016-06-29 at 14:37 +0200, Hans de Goede wrote:
> In order to implement get_work_dim() the driver may need to know the
> clEnqueueNDRangeKernel() work_dim parameter, so pass it to the
> driver.

The workdim info is passed as the first implicit argument (after
explicit kernel arguments, see invocation.cpp:600 ). Can you use that?
Both radeonsi and r600 already do.

Jan

> 
> Signed-off-by: Hans de Goede <hdegoede at redhat.com>
> ---
> Changes in v2:
> -No changes
> ---
>  src/gallium/include/pipe/p_state.h                | 7 +++++++
>  src/gallium/state_trackers/clover/core/kernel.cpp | 1 +
>  2 files changed, 8 insertions(+)
> 
> diff --git a/src/gallium/include/pipe/p_state.h
> b/src/gallium/include/pipe/p_state.h
> index 1543e90..78f5374 100644
> --- a/src/gallium/include/pipe/p_state.h
> +++ b/src/gallium/include/pipe/p_state.h
> @@ -746,6 +746,13 @@ struct pipe_grid_info
>     void *input;
>  
>     /**
> +    * Grid number of dimensions, 1-3, e.g. the work_dim parameter
> passed to
> +    * clEnqueueNDRangeKernel. Note block[] and grid[] must be padded
> with
> +    * 1 for non-used dimensions.
> +    */
> +   uint work_dim;
> +
> +   /**
>      * Determine the layout of the working block (in thread units) to
> be used.
>      */
>     uint block[3];
> diff --git a/src/gallium/state_trackers/clover/core/kernel.cpp
> b/src/gallium/state_trackers/clover/core/kernel.cpp
> index 9231462..d9bda6c 100644
> --- a/src/gallium/state_trackers/clover/core/kernel.cpp
> +++ b/src/gallium/state_trackers/clover/core/kernel.cpp
> @@ -76,6 +76,7 @@ kernel::launch(command_queue &q,
>                                exec.g_buffers.data(),
> g_handles.data());
>  
>     // Fill information for the launch_grid() call.
> +   info.work_dim = grid_size.size();
>     copy(pad_vector(q, block_size, 1), info.block);
>     copy(pad_vector(q, reduced_grid_size, 1), info.grid);
>     info.pc = find(name_equals(_name), m.syms).offset;
-- 

Jan Vesely <jan.vesely at rutgers.edu>
-------------- next part --------------
A non-text attachment was scrubbed...
Name: signature.asc
Type: application/pgp-signature
Size: 819 bytes
Desc: This is a digitally signed message part
URL: <https://lists.freedesktop.org/archives/mesa-dev/attachments/20160629/97a8a023/attachment.sig>