<commit_before>{-# LANGUAGE ForeignFunctionInterface, GADTs #-}
-- |
-- Module      : Foreign.OpenCL.Bindings.Kernel
-- Copyright   : (c) 2011, <NAME>
-- License     : BSD3
-- 
-- Maintainer  : <NAME> <<EMAIL>>
-- Stability   : experimental
-- Portability : non-portable (GHC extensions)
--
-- 
-- OpenCL kernel creation, invocation and scheduling.

module Foreign.OpenCL.Bindings.Kernel (
   createKernel,

   kernelContext, kernelFunctionName, kernelNumArgs,
   kernelWorkGroupSize, kernelCompileWorkGroupSize, kernelLocalMemSize,
   kernelPreferredWorkGroupSizeMultiple, kernelPrivateMemSize,

   enqueueNDRangeKernel, enqueueTask,

   KernelArg(..), setKernelArg, setKernelArgs
  ) where

#include <CL/cl.h>

import Control.Monad

import Foreign
import Foreign.C.String
import Foreign.C.Types

{# import Foreign.OpenCL.Bindings.Internal.Types #}
import Foreign.OpenCL.Bindings.Internal.Finalizers
import Foreign.OpenCL.Bindings.Internal.Error
import Foreign.OpenCL.Bindings.Internal.Util

-- | Create a program from a string containing the source code
--
createKernel :: Program -- ^The program that contains the kernel code
              -> String -- ^The name of the kernel (as written in the program source)
              -> IO Kernel -- ^The newly created kernel
createKernel prog name =
   withForeignPtr prog $ \prog_ptr ->
   withCString name $ \cstr ->
   alloca $ \ep -> do
      kernel <- {# call unsafe clCreateKernel #} prog_ptr cstr ep
      checkClError_ "clCreateKernel" =<< peek ep
      attachFinalizer kernel

-- | Enqueues a command to execute a given kernel on a device. See section 5.8 in the OpenCL 1.1 specification
enqueueNDRangeKernel :: CommandQueue 
                     -> Kernel
                     -> [ClSize] -- ^ Global work offsets
                     -> [ClSize] -- ^ Global work sizes
                     -> [ClSize] -- ^ Local work sizes
                     -> [Event] 
                     -> IO Event
enqueueNDRangeKernel cq k globalWorkOffsets globalWorkSizes localWorkSizes waitEvs =
    withForeignPtr cq $ \queue ->
    withForeignPtr k $ \kernel ->
    withArrayNull globalWorkOffsets $ \globalWorkOffsetPtr ->
    withArrayNull globalWorkSizes $ \globalWorkSizePtr ->
    withArrayNull localWorkSizes $ \localWorkSizePtr ->
    withForeignPtrs waitEvs $ \event_ptrs ->
    withArrayNullLen event_ptrs $ \n event_array ->
    alloca $ \eventPtr ->
    do checkClError_ "clEnqueueNDRangeKernel" =<< 
         {# call unsafe clEnqueueNDRangeKernel #} 
                   queue kernel workDim
                   globalWorkOffsetPtr
                   globalWorkSizePtr
                   localWorkSizePtr
                   (fromIntegral n)
                   event_array
                   eventPtr
       attachFinalizer =<< peek eventPtr
  where workDim = fromIntegral . maximum $ map length [globalWorkOffsets, globalWorkSizes, localWorkSizes]

data KernelArg where
  MObjArg :: MemObject a -> KernelArg
  LocalArrayArg :: Storable a => a -> Int -> KernelArg
  VArg :: Storable a => a -> KernelArg
  StructArg :: Storable a => [a] -> KernelArg

-- | Invoking @setKernelArg krn n arg@ sets argument @n@ of the kernel @krn@
setKernelArg :: Kernel -> Int -> KernelArg -> IO ()
setKernelArg kernel n param =
  withForeignPtr kernel $ \k ->
  withPtr param $ \param_ptr -> do
    err <- {# call unsafe clSetKernelArg #} k (fromIntegral n) (size param) param_ptr
    case toEnum $ fromIntegral err of
      InvalidArgSize  -> error $ "ClInvalidArgSize occurred in call to: clSetKernelArg. Argument #"
                                 ++ show n ++ " was set to size " ++ show (size param)
      InvalidArgIndex -> error $ "ClInvalidArgIndex occurred in call to: clSetKernelArg, when setting argument #"
                                 ++ show n
      _ -> checkClError_ "clSetKernelArg" err
      where size (MObjArg mobj) = fromIntegral $ sizeOf (memobjPtr mobj)
            size (VArg v) = fromIntegral $ sizeOf v
            size (StructArg xs) = fromIntegral . sum $ map sizeOf xs
            size (LocalArrayArg x m) = fromIntegral $ m * sizeOf x        <commit_msg>Get rid of some type defaulting warnings<commit_after>{-# LANGUAGE ForeignFunctionInterface, GADTs #-}
-- |
-- Module      : Foreign.OpenCL.Bindings.Kernel
-- Copyright   : (c) 2011, <NAME>
-- License     : BSD3
-- 
-- Maintainer  : <NAME> <<EMAIL>>
-- Stability   : experimental
-- Portability : non-portable (GHC extensions)
--
-- 
-- OpenCL kernel creation, invocation and scheduling.

module Foreign.OpenCL.Bindings.Kernel (
   createKernel,

   kernelContext, kernelFunctionName, kernelNumArgs,
   kernelWorkGroupSize, kernelCompileWorkGroupSize, kernelLocalMemSize,
   kernelPreferredWorkGroupSizeMultiple, kernelPrivateMemSize,

   enqueueNDRangeKernel, enqueueTask,

   KernelArg(..), setKernelArg, setKernelArgs
  ) where

#include <CL/cl.h>

import Control.Monad

import Foreign
import Foreign.C.String
import Foreign.C.Types

{# import Foreign.OpenCL.Bindings.Internal.Types #}
import Foreign.OpenCL.Bindings.Internal.Finalizers
import Foreign.OpenCL.Bindings.Internal.Error
import Foreign.OpenCL.Bindings.Internal.Util

-- | Create a program from a string containing the source code
--
createKernel :: Program -- ^The program that contains the kernel code
              -> String -- ^The name of the kernel (as written in the program source)
              -> IO Kernel -- ^The newly created kernel
createKernel prog name =
   withForeignPtr prog $ \prog_ptr ->
   withCString name $ \cstr ->
   alloca $ \ep -> do
      kernel <- {# call unsafe clCreateKernel #} prog_ptr cstr ep
      checkClError_ "clCreateKernel" =<< peek ep
      attachFinalizer kernel

-- | Enqueues a command to execute a given kernel on a device. See section 5.8 in the OpenCL 1.1 specification
enqueueNDRangeKernel :: CommandQueue 
                     -> Kernel
                     -> [ClSize] -- ^ Global work offsets
                     -> [ClSize] -- ^ Global work sizes
                     -> [ClSize] -- ^ Local work sizes
                     -> [Event] 
                     -> IO Event
enqueueNDRangeKernel cq k globalWorkOffsets globalWorkSizes localWorkSizes waitEvs =
    withForeignPtr cq $ \queue ->
    withForeignPtr k $ \kernel ->
    withArrayNull globalWorkOffsets $ \globalWorkOffsetPtr ->
    withArrayNull globalWorkSizes $ \globalWorkSizePtr ->
    withArrayNull localWorkSizes $ \localWorkSizePtr ->
    withForeignPtrs waitEvs $ \event_ptrs ->
    withArrayNullLen event_ptrs $ \n event_array ->
    alloca $ \eventPtr ->
    do checkClError_ "clEnqueueNDRangeKernel" =<< 
         {# call unsafe clEnqueueNDRangeKernel #} 
                   queue kernel workDim
                   globalWorkOffsetPtr
                   globalWorkSizePtr
                   localWorkSizePtr
                   (fromIntegral n)
                   event_array
                   eventPtr
       attachFinalizer =<< peek eventPtr
  where workDim = fromIntegral . maximum $ map length [globalWorkOffsets, globalWorkSizes, localWorkSizes]

data KernelArg where
  MObjArg :: MemObject a -> KernelArg
  LocalArrayArg :: Storable a => a -> Int -> KernelArg
  VArg :: Storable a => a -> KernelArg
  StructArg :: Storable a => [a] -> KernelArg

-- | Invoking @setKernelArg krn n arg@ sets argument @n@ of the kernel @krn@
setKernelArg :: Kernel -> Int -> KernelArg -> IO ()
setKernelArg kernel n param =
  withForeignPtr kernel $ \k ->
  withPtr param $ \param_ptr -> do
    err <- {# call unsafe clSetKernelArg #} k (fromIntegral n) (size param) param_ptr
    case toEnum $ fromIntegral err of
      InvalidArgSize  -> error $ "ClInvalidArgSize occurred in call to: clSetKernelArg. Argument #"
                                 ++ show n ++ " was set to size " ++ show (size param)
      InvalidArgIndex -> error $ "ClInvalidArgIndex occurred in call to: clSetKernelArg, when setting argument #"
                                 ++ show n
      _ -> checkClError_ "clSetKernelArg" err
      where size :: KernelArg -> CULong
            size (MObjArg mobj) = fromIntegral $ sizeOf (memobjPtr mobj)
            size (VArg v) = fromIntegral $ sizeOf v
            size (StructArg xs) = fromIntegral . sum $ map sizeOf xs
            size (LocalArrayArg x m) = fromIntegral $ m * sizeOf x        
{'fraction_non_alphanumeric': 0.08818214672930973, 'fraction_numerical': 0.0021684134441633538, 'mean_word_length': 2.457725947521866, 'pattern_counts': {'":': 0, '<': 29, '<?xml version=': 0, '>': 84, 'https://': 0, 'lorem ipsum': 0, 'www.': 0, 'xml': 0}, 'pii_count': 0, 'substrings_counts': 0, 'word_list_counts': {'cursed_substrings.json': 6, 'profanity_word_list.json': 0, 'sexual_word_list.json': 0, 'zh_pornsignals.json': 0}}
{'dir': 'git-commits-cleaned', 'id': '6867818', 'n_tokens_mistral': 2377, 'n_tokens_neox': 2294, 'n_words': 709}