Fixes #86

This fixes the previous commit.

These functions expect all index parameters to have the same type, so we
need to cast 0 to the IndexType, otherwise the function lookup will fail
when IndexType is not int.

Added __device__ flag to the Shared union constructors in CudaReductionKernel.h and CudaScanKernel.h to make clang happy

Addresses #86

Also fixed the header in Preconditioner.h

MPI refactoring

See merge request !101

It does not make sense to print a distributed array like this, because
it may contain ghost elements. Users should examine the local array
view manually.

Replaced send/receive for Array(,View) and mpiSend/mpiReceive for String with a general implementation in the MPI namespace

Also added analogous functions: MPI::sendrecv, MPI::bcast.

- removed wrapper functions: AllGroup, NullGroup, NullRequest
- added MPI_COMM_WORLD and other handles to MPI/DummyDefs.h
- renamed getCommunicationGroup to getCommunicator in all data structures
- improved naming to match the MPI terminology: communicator instead of group

CUDA kernels should not ever work with distributed data structures, they
should always get the underlying *local* data structure.

- The "NullGroup" should not be used even when built without MPI,
  otherwise the behaviour is very bug-prone because "NullGroup" usage is
  not caught and changing the build type leads to a different semantics.
- "AllGroup" is not a good default value for the parameters, considering
  that the class attributes are initialized to "NullGroup".

Buffering with a small value is very slow when profiling anything in nvpp.

Amends e5fc6a96

Scan refactoring

Closes #87

See merge request !100

- structs from HorizontalOperations.h reimplemented as function objects in
  Functional.h
- repetitive function definitions generated using macros
- added new operators: % (modulus) and ^ (xor)