ft_1d.hs

{-# LANGUAGE ScopedTypeVariables #-}
import Control.Parallel.OpenCL
import Foreign( castPtr, nullPtr, sizeOf )
import Foreign.C.Types( CFloat,CDouble )
import Foreign.Marshal.Array( newArray, peekArray )
import System.IO 
import System.Exit
import qualified Control.Exception as E
import Control.Monad
import Data.List

--error handling
onError :: String -> IOError -> IO String
onError filename error = do
  hPutStrLn stderr $ "File not found: " ++ filename
  exitWith(ExitFailure 1)

--read from file
programSourceFromFile :: IO String
programSourceFromFile = do let clprogramName = "ft_1d.cl"
                           E.catch (readFile clprogramName)
                                 (onError clprogramName)

-- transform String to [CDouble]
toNum :: String -> [CDouble]
toNum = foldr toNumber [] . words
      where toNumber xs number = read xs:number

-- transform [CDouble] to String
toStrings :: [CDouble] -> String
toStrings = unlines . foldr toStr [] 
    where toStr xs number = show xs:number

-- read input value from File
numberFromFile :: IO String
numberFromFile = do let cltransformSource = "input.txt"
                    E.catch (readFile cltransformSource)
                          (onError cltransformSource)

getDeviceIDs :: CLDeviceType -> CLPlatformID -> IO [CLDeviceID]
getDeviceIDs devtype platforms = do
  dev <- clGetDeviceIDs platforms devtype
  return dev

executeCLMain :: IO ()
executeCLMain = do
  -- initialize OpenCL
  platforms <- clGetPlatformIDs
  putStrLn $ show platforms
  mapM_ clDevice [platforms]

clDevice :: [CLPlatformID] -> IO ()
clDevice platform  
  = forM_ platform $ \platform -> do
    devs <- clGetDeviceIDs platform CL_DEVICE_TYPE_ALL
    mapM_ clDeviceGPU devs

-- execute OpenCL Kernel Program on GPU Type Device
clDeviceGPU :: CLDeviceID -> IO ()    
clDeviceGPU dev = do
  devtype <- clGetDeviceType dev
  putStrLn $ show devtype 
  when (show devtype == show [CL_DEVICE_TYPE_GPU])
           $ mapM_ execCL [dev]

execCL :: CLDeviceID -> IO ()
execCL dev = do
  context <- clCreateContext [] [dev] print
  q <- clCreateCommandQueue context dev []
 
  -- Compile and get binaries
  clsource <- programSourceFromFile
  program <- clCreateProgramWithSource context clsource
  clBuildProgram program [dev] ""
  bins <- clGetProgramBinaries program
  
  -- Create kernel from binaries
  (program2,_) <- clCreateProgramWithBinary context [dev] bins 
  let clkernelBuildOption = ""
      clkernelName = "arrayft_1d"
  clBuildProgram program2 [dev] clkernelBuildOption
  kernel <- clCreateKernel program2 clkernelName
  -- read transform source
  insource <- numberFromFile
  
  -- Initialize parameters
  let original = toNum insource -- read from file [CDouble] value
      elemSize = sizeOf (0 :: CDouble) -- or CFloat
      vecSize = elemSize * length original
      
  putStrLn $ "Original array = \n" ++ toStrings original
  input <- newArray original
 
  mem_in <- clCreateBuffer context [CL_MEM_READ_ONLY, CL_MEM_COPY_HOST_PTR] (vecSize, castPtr input)
  mem_out <- clCreateBuffer context [CL_MEM_WRITE_ONLY] (vecSize, nullPtr)
                    
  clSetKernelArgSto kernel 0 mem_in
  clSetKernelArgSto kernel 1 mem_out
  
  {- Execute Kernel
   * clEnqueueNDRangeKernel 
     1. commandqueue 
     2. clkernel
     3. global work size
     4. local work size (perhaps best for performance => [](NULL) )
     5. clevent (always NULL)
  -}
  eventExec <- clEnqueueNDRangeKernel q kernel [length original] [] []

  -- Get Result
  eventRead <- clEnqueueReadBuffer q mem_out True 0 vecSize (castPtr input) [eventExec]
  result <- peekArray (length original) input
  putStrLn $ "Result array   = \n" ++ toStrings result
  return ()

main :: IO ()
main = executeCLMain