Add cooperative kernel launch API

[LutzCle]

Pascal and newer devices support cooperative groups. These groups enable kernels to, e.g., globally synchronize the grid without terminating the kernel.

A special function cuLaunchCooperativeKernel was added to CUDA 9.0 to launch a cooperative group. This PR adds this launch API to RustaCUDA.

Example:

unsafe { launch_cooperative!( kernel<<<grid, block>>>() ).unwrap() };

See the CUDA documentation and this Nvidia Developer blog post for more information.