Source

OpenCL Raw / Walkthrough.hs

{-# OPTIONS_GHC -framework OpenCL #-}
import Foreign.Ptr
import Foreign.Storable
import System.OpenCL.Raw.V10
import Control.Monad.Trans
import Control.Monad
import Foreign.Marshal.Alloc


newtype EitherT m l r = EitherT {runEitherT :: m (Either l r)}
instance Monad m => Monad (EitherT m l) where
    return = EitherT . return . Right
    (EitherT me) >>= f = EitherT $ me >>=  \e -> 
        case e of
            Right a  -> runEitherT $ f a
            Left  a  -> return $ Left a

instance MonadIO m => MonadIO (EitherT m l) where
    liftIO = EitherT . liftM Right . liftIO

mbToEither Nothing = Right ()
mbToEither (Just err) = Left err

defaultPlatform = nullPtr
main = do
    res <- runEitherT $ do
        devices <- EitherT $ clGetDeviceIDs nullPtr clDeviceTypeAll 100 
        let callback _ _ _ _ = return ()
        ctx <- EitherT $ clCreateContext [] devices callback nullPtr
        let device = head devices
        queue <- EitherT $ clCreateCommandQueue ctx device (CommandQueueProperties 0)
        prog <- EitherT $ clCreateProgramWithSource ctx programSrc 
        let progCB _ _ = return ()
        EitherT $ liftM mbToEither $ clBuildProgram prog [device]  "" progCB nullPtr 
        kern <- EitherT $ clCreateKernel prog "square"
        inp   <- liftIO malloc
        outp  <- liftIO malloc
        inpB  <- EitherT $ clCreateBuffer ctx (MemFlags $ 8 + 4)  (fromIntegral $ sizeOf inp) (castPtr (inp :: Ptr Double))
        outpB <- EitherT $ clCreateBuffer ctx (MemFlags $ 8 + 2) (fromIntegral $ sizeOf outp) (castPtr (outp :: Ptr Double))
        inpBStorage <- liftIO malloc
        outBStorage <- liftIO malloc
        liftIO $ poke inpBStorage inpB >> poke outBStorage outpB
        EitherT $ liftM mbToEither $             poke inp 2.3 >> clSetKernelArg kern 0 (fromIntegral $ sizeOf (undefined::Ptr ())) (castPtr inpBStorage)
        EitherT $ liftM mbToEither $                             clSetKernelArg kern 1 (fromIntegral $ sizeOf (undefined::Ptr ())) (castPtr outBStorage)
        EitherT $ liftM mbToEither $ alloca $ \p  -> poke p 1 >> clSetKernelArg kern 2 (fromIntegral $ sizeOf (undefined::CLint)) (castPtr (p :: Ptr CLint))
        event <- EitherT $ clEnqueueTask queue kern []
        EitherT $ liftM mbToEither $ clWaitForEvents [event]
        liftIO $ peek outp



    print res
    

programSrc :: String
programSrc = "\
\__kernel square(                                                        \n \
\   __global double* input,                                              \n \
\   __global double* output,                                             \n \
\   const unsigned int count)                                           \n \
\{                                                                      \n \
\   int i = get_global_id(0);                                           \n \
\   if(i < count)                                                       \n \
\       output[i] = input[i] * input[i]    ;                                \n \
\}                                                                      \n \
\"