 #define ASSERT_SUCCESS(res) ASSERT_EQ(cudaSuccess, (res));
 #define ASSERT_FAILURE(res) ASSERT_NE(cudaSuccess, (res));
 
+// for testing hemi::launch()
 struct Kernel {
 	template <typename... Arguments>
 	HEMI_DEV_CALLABLE_MEMBER void operator()(int *count, int *bdim, int *gdim, Arguments... args) {

	ep.setSharedMemBytes(1000000);
	hemi::launch(ep, kernel, dCount, dBdim, dGdim);
	ASSERT_FAILURE(checkCudaErrors());
}
	}
};

// for testing hemi::cudaLaunch()
template <typename... Arguments>
HEMI_LAUNCHABLE void KernelFunc(int *count, int *bdim, int *gdim, Arguments... args) {
	Kernel k;
	k(count, bdim, gdim, args...);
}

class LaunchTest : public ::testing::Test {
protected:
  virtual void SetUp() {
  	cudaMalloc(&dCount, sizeof(int));
    cudaMalloc(&dBdim, sizeof(int));
