cuArraysCopy.cu File Reference

Utilities for copying/converting images to different format. More...

#include "cuArrays.h"
#include "cudaUtil.h"
#include "cudaError.h"
#include "float2.h"

Functions
__global__ void	cuArraysCopyToBatch_kernel (const float2 *imageIn, const int inNX, const int inNY, float2 *imageOut, const int outNX, const int outNY, const int nImagesX, const int nImagesY, const int strideX, const int strideY)
void	cuArraysCopyToBatch (cuArrays< float2 > *image1, cuArrays< float2 > *image2, int strideH, int strideW, cudaStream_t stream)
	Copy a chunk into a batch of chips for a given stride.
template<typename T_in, typename T_out>
__global__ void	cuArraysCopyToBatchWithOffset_kernel (const T_in *imageIn, const int inNY, T_out *imageOut, const int outNX, const int outNY, const int nImages, const int *offsetX, const int *offsetY)
void	cuArraysCopyToBatchWithOffset (cuArrays< float2 > *image1, const int lda1, cuArrays< float2 > *image2, const int *offsetH, const int *offsetW, cudaStream_t stream)
	Copy a chunk into a batch of chips with varying offsets/strides.
__global__ void	cuArraysCopyToBatchAbsWithOffset_kernel (const float2 *imageIn, const int inNY, float2 *imageOut, const int outNX, const int outNY, const int nImages, const int *offsetX, const int *offsetY)
void	cuArraysCopyToBatchAbsWithOffset (cuArrays< float2 > *image1, const int lda1, cuArrays< float2 > *image2, const int *offsetH, const int *offsetW, cudaStream_t stream)
	Copy a chunk into a batch of chips with varying offsets/strides.
void	cuArraysCopyToBatchWithOffsetR2C (cuArrays< float > *image1, const int lda1, cuArrays< float2 > *image2, const int *offsetH, const int *offsetW, cudaStream_t stream)
	Copy a chunk into a batch of chips with varying offsets/strides.
__global__ void	cuArraysCopyC2R_kernel (const float2 *imageIn, const int inNX, const int inNY, float *imageOut, const int outNX, const int outNY, const int nImagesX, const int nImagesY, const int strideX, const int strideY, const float factor)
void	cuArraysCopyC2R (cuArrays< float2 > *image1, cuArrays< float > *image2, int strideH, int strideW, cudaStream_t stream)
	Copy a chunk into a batch of chips with varying offsets/strides.
template<typename T>
__global__ void	cuArraysCopyExtractVaryingOffset (const T *imageIn, const int inNX, const int inNY, T *imageOut, const int outNX, const int outNY, const int nImages, const int2 *offsets)
template<typename T>
void	cuArraysCopyExtract (cuArrays< T > *imagesIn, cuArrays< T > *imagesOut, cuArrays< int2 > *offsets, cudaStream_t stream)
	Copy a tile of images to another image, with starting pixels offsets.
template void	cuArraysCopyExtract (cuArrays< float > *in, cuArrays< float > *out, cuArrays< int2 > *offsets, cudaStream_t)
template void	cuArraysCopyExtract (cuArrays< float2 > *in, cuArrays< float2 > *out, cuArrays< int2 > *offsets, cudaStream_t)
__global__ void	cuArraysCopyExtractVaryingOffsetCorr (const float *imageIn, const int inNX, const int inNY, float *imageOut, const int outNX, const int outNY, int *imageValid, const int nImages, const int2 *maxloc)
void	cuArraysCopyExtractCorr (cuArrays< float > *imagesIn, cuArrays< float > *imagesOut, cuArrays< int > *imagesValid, cuArrays< int2 > *maxloc, cudaStream_t stream)
	copy a tile of images to another image, with starting pixels offsets accounting for boundary
template<typename T>
__global__ void	cuArraysCopyExtractFixedOffset (const T *imageIn, const int inNX, const int inNY, T *imageOut, const int outNX, const int outNY, const int nImages, const int offsetX, const int offsetY)
__global__ void	cuArraysCopyExtractFixedOffset (const float2 *imageIn, const int inNX, const int inNY, float *imageOut, const int outNX, const int outNY, const int nImages, const int offsetX, const int offsetY)
template<typename T_in, typename T_out>
void	cuArraysCopyExtract (cuArrays< T_in > *imagesIn, cuArrays< T_out > *imagesOut, int2 offset, cudaStream_t stream)
	copy/extract images from a large size to a smaller size from the location (offsetX, offsetY)
template void	cuArraysCopyExtract (cuArrays< float > *in, cuArrays< float > *out, int2 offset, cudaStream_t)
template void	cuArraysCopyExtract (cuArrays< float2 > *in, cuArrays< float > *out, int2 offset, cudaStream_t)
template void	cuArraysCopyExtract (cuArrays< float2 > *in, cuArrays< float2 > *out, int2 offset, cudaStream_t)
template void	cuArraysCopyExtract (cuArrays< float3 > *in, cuArrays< float3 > *out, int2 offset, cudaStream_t)
template<typename T>
__global__ void	cuArraysCopyInsert_kernel (const T *imageIn, const int inNX, const int inNY, T *imageOut, const int outNY, const int offsetX, const int offsetY)
template<typename T>
void	cuArraysCopyInsert (cuArrays< T > *imageIn, cuArrays< T > *imageOut, int offsetX, int offsetY, cudaStream_t stream)
	copy/insert images from a smaller size to a larger size from the location (offsetX, offsetY)
template void	cuArraysCopyInsert (cuArrays< float2 > *in, cuArrays< float2 > *out, int offX, int offY, cudaStream_t)
template void	cuArraysCopyInsert (cuArrays< float3 > *in, cuArrays< float3 > *out, int offX, int offY, cudaStream_t)
template void	cuArraysCopyInsert (cuArrays< float > *in, cuArrays< float > *out, int offX, int offY, cudaStream_t)
template void	cuArraysCopyInsert (cuArrays< int > *in, cuArrays< int > *out, int offX, int offY, cudaStream_t)
template<typename T_in, typename T_out>
__global__ void	cuArraysCopyPadded_kernel (T_in *imageIn, int inNX, int inNY, int sizeIn, T_out *imageOut, int outNX, int outNY, int sizeOut, int nImages)
template<typename T_in, typename T_out>
void	cuArraysCopyPadded (cuArrays< T_in > *imageIn, cuArrays< T_out > *imageOut, cudaStream_t stream)
	copy images from a smaller size to a larger size while padding 0 for extra elements
template void	cuArraysCopyPadded (cuArrays< float > *imageIn, cuArrays< float > *imageOut, cudaStream_t)
template void	cuArraysCopyPadded (cuArrays< float > *imageIn, cuArrays< float2 > *imageOut, cudaStream_t)
template void	cuArraysCopyPadded (cuArrays< float2 > *imageIn, cuArrays< float2 > *imageOut, cudaStream_t)
__global__ void	cuArraysSetConstant_kernel (float *image, int size, float value)
void	cuArraysSetConstant (cuArrays< float > *imageIn, float value, cudaStream_t stream)
	Set real images to a constant value.
__global__ void	cuArraysR2C_kernel (float *image1, float2 *image2, int size)
void	cuArraysR2C (cuArrays< float > *image1, cuArrays< float2 > *image2, cudaStream_t stream)
	Convert real images to complex images (set imaginary parts to 0)
__global__ void	cuArraysC2R_kernel (float2 *image1, float *image2, int size)
void	cuArraysC2R (cuArrays< float2 > *image1, cuArrays< float > *image2, cudaStream_t stream)
	Take real part of complex images.
__global__ void	cuArraysAbs_kernel (float2 *image1, float *image2, int size)
void	cuArraysAbs (cuArrays< float2 > *image1, cuArrays< float > *image2, cudaStream_t stream)
	Obtain abs (amplitudes) of complex images.

Detailed Description

Utilities for copying/converting images to different format.

All methods are declared in cuAmpcorUtil.h cudaArraysCopyToBatch to extract a batch of windows from the raw image various implementations include:

1. fixed or varying offsets, as start pixels for windows
2. complex to complex, usually
3. complex to (amplitude,0), for TOPS
4. real to complex, for real images cuArraysCopyExtract to extract(shrink in size) from a batch of windows to another batch overloaded for different data types cuArraysCopyInsert to insert a batch of windows (smaller in size) to another batch overloaded for different data types cuArraysCopyPadded to insert a batch of windows to another batch while padding 0s for rest elements used for fft oversampling see also cuArraysPadding.cu for other zero-padding utilities cuArraysR2C cuArraysC2R cuArraysAbs to convert between different data types

All methods are declared in cuAmpcorUtil.h cudaArraysCopyToBatch to extract a batch of windows from the raw image various implementations include:

1. fixed or varying offsets, as start pixels for windows
2. complex to complex, usually
3. complex to (amplitude,0), for TOPS
4. real to complex, for real images cuArraysCopyExtract to extract(shrink in size) from a batch of windows to another batch overloaded for different data types cuArraysCopyInsert to insert a batch of windows (smaller in size) to another batch overloaded for different data types cuArraysCopyPadded to insert a batch of windows to another batch while padding 0s for rest elements used for fft oversampling see also cuArraysPadding.cu for other zero-padding utilities cuArraysAbs to convert between different data types

Function Documentation

cuArraysAbs()

void cuArraysAbs	(	cuArrays< float2 > *	image1,
		cuArrays< float > *	image2,
		cudaStream_t	stream )

Obtain abs (amplitudes) of complex images.

Parameters

[in]	image1	input images
[out]	image2	output images

cuArraysC2R()

void cuArraysC2R	(	cuArrays< float2 > *	image1,
		cuArrays< float > *	image2,
		cudaStream_t	stream )

Take real part of complex images.

Parameters

[in]	image1	input images
[out]	image2	output images

cuArraysCopyC2R()

void cuArraysCopyC2R	(	cuArrays< float2 > *	image1,
		cuArrays< float > *	image2,
		int	strideH,
		int	strideW,
		cudaStream_t	stream )

Copy a chunk into a batch of chips with varying offsets/strides.

Note

similar to cuArraysCopyToBatchWithOffset, but take amplitudes instead

Parameters

image1	Input image as a large chunk
image2	Output images as a batch of chips
strideH	offsets along height to extract chips
strideW	offsets along width to extract chips
stream	cudaStream

cuArraysCopyExtract()

template<typename T>

void cuArraysCopyExtract	(	cuArrays< T > *	imagesIn,
		cuArrays< T > *	imagesOut,
		cuArrays< int2 > *	offsets,
		cudaStream_t	stream )

Copy a tile of images to another image, with starting pixels offsets.

Parameters

[in]	imageIn	input images of dimension nImages*inNX*inNY
[out]	imageOut	output images of dimension nImages*outNX*outNY
[in]	offsets,varying	offsets for extraction

cuArraysCopyExtractCorr()

void cuArraysCopyExtractCorr	(	cuArrays< float > *	imagesIn,
		cuArrays< float > *	imagesOut,
		cuArrays< int > *	imagesValid,
		cuArrays< int2 > *	maxloc,
		cudaStream_t	stream )

copy a tile of images to another image, with starting pixels offsets accounting for boundary

Parameters

[in]	imageIn	input images
[out]	imageOut	output images of dimension nImages*outNX*outNY

cuArraysCopyToBatch()

void cuArraysCopyToBatch	(	cuArrays< float2 > *	image1,
		cuArrays< float2 > *	image2,
		int	strideH,
		int	strideW,
		cudaStream_t	stream )

Copy a chunk into a batch of chips for a given stride.

Note

used to extract chips from a raw image

Parameters

image1	Input image as a large chunk
image2	Output images as a batch of chips
strideH	stride along height to extract chips
strideW	stride along width to extract chips
stream	cudaStream

cuArraysCopyToBatchAbsWithOffset()

void cuArraysCopyToBatchAbsWithOffset	(	cuArrays< float2 > *	image1,
		const int	lda1,
		cuArrays< float2 > *	image2,
		const int *	offsetH,
		const int *	offsetW,
		cudaStream_t	stream )

Copy a chunk into a batch of chips with varying offsets/strides.

Note

similar to cuArraysCopyToBatchWithOffset, but take amplitudes instead

Parameters

image1	Input image as a large chunk
lda1	the leading dimension of image1, usually, its width inNY
image2	Output images as a batch of chips
strideH	(varying) offsets along height to extract chips
strideW	(varying) offsets along width to extract chips
stream	cudaStream

cuArraysCopyToBatchWithOffset()

void cuArraysCopyToBatchWithOffset	(	cuArrays< float2 > *	image1,
		const int	lda1,
		cuArrays< float2 > *	image2,
		const int *	offsetH,
		const int *	offsetW,
		cudaStream_t	stream )

Copy a chunk into a batch of chips with varying offsets/strides.

Note

used to extract chips from a raw secondary image with varying offsets

Parameters

image1	Input image as a large chunk
lda1	the leading dimension of image1, usually, its width inNY
image2	Output images as a batch of chips
strideH	(varying) offsets along height to extract chips
strideW	(varying) offsets along width to extract chips
stream	cudaStream

cuArraysCopyToBatchWithOffsetR2C()

void cuArraysCopyToBatchWithOffsetR2C	(	cuArrays< float > *	image1,
		const int	lda1,
		cuArrays< float2 > *	image2,
		const int *	offsetH,
		const int *	offsetW,
		cudaStream_t	stream )

Copy a chunk into a batch of chips with varying offsets/strides.

Note

used to load real images

Parameters

image1	Input image as a large chunk
lda1	the leading dimension of image1, usually, its width inNY
image2	Output images as a batch of chips
strideH	(varying) offsets along height to extract chips
strideW	(varying) offsets along width to extract chips
stream	cudaStream

cuArraysR2C()

void cuArraysR2C	(	cuArrays< float > *	image1,
		cuArrays< float2 > *	image2,
		cudaStream_t	stream )

Convert real images to complex images (set imaginary parts to 0)

Parameters

[in]	image1	input images
[out]	image2	output images

cuArraysSetConstant()

void cuArraysSetConstant	(	cuArrays< float > *	imageIn,
		float	value,
		cudaStream_t	stream )

Set real images to a constant value.

Note

use setZero if value=0 because cudaMemset is faster