pub struct PartitionView<T> { /* private fields */ }Expand description
A tiling strategy over a TensorView.
PartitionView divides a tensor into tiles of a specified shape, enabling efficient GPU processing with shared memory optimization.
§Type Parameters
T- Element type of the underlying tensor
§cuda-tile-behavior.md References
- Falsification test #36: Tile count calculation is correct
- Falsification test #37: Tile iteration covers all elements
- Falsification test #38: Edge tiles are handled correctly
Implementations§
Source§impl<T> PartitionView<T>
impl<T> PartitionView<T>
Sourcepub fn new(tensor: TensorView<T>, tile_shape: [usize; 4]) -> Self
pub fn new(tensor: TensorView<T>, tile_shape: [usize; 4]) -> Self
Sourcepub fn new_power_of_two(tensor: TensorView<T>, tile_log2: [usize; 4]) -> Self
pub fn new_power_of_two(tensor: TensorView<T>, tile_log2: [usize; 4]) -> Self
Create a PartitionView with power-of-two tile sizes.
This is recommended for GPU compute as it enables efficient memory coalescing and avoids bank conflicts.
§Arguments
tensor- The tensor to partitiontile_log2- Log2 of tile size for each dimension
§cuda-tile-behavior.md References
- Falsification test #1: Power-of-two tiles improve GPU occupancy
Sourcepub fn new_2d(tensor: TensorView<T>, tile_rows: usize, tile_cols: usize) -> Self
pub fn new_2d(tensor: TensorView<T>, tile_rows: usize, tile_cols: usize) -> Self
Create a PartitionView with 2D tiles (for matrix operations).
§Arguments
tensor- The tensor to partitiontile_rows- Number of rows per tiletile_cols- Number of columns per tile
Sourcepub fn tensor(&self) -> &TensorView<T>
pub fn tensor(&self) -> &TensorView<T>
Get the underlying tensor.
Sourcepub fn tile_shape(&self) -> &[usize; 4]
pub fn tile_shape(&self) -> &[usize; 4]
Get the tile shape.
Sourcepub fn tile_count(&self) -> [usize; 4]
pub fn tile_count(&self) -> [usize; 4]
Get the number of tiles in each dimension.
§cuda-tile-behavior.md References
- Falsification test #36: Tile count calculation is correct
Sourcepub fn total_tiles(&self) -> usize
pub fn total_tiles(&self) -> usize
Get the total number of tiles.
Sourcepub fn get_tile_view(&self, tile_idx: [usize; 4]) -> Option<TensorView<T>>
pub fn get_tile_view(&self, tile_idx: [usize; 4]) -> Option<TensorView<T>>
Sourcepub fn iter_tiles(&self) -> TileIterator<'_, T>
pub fn iter_tiles(&self) -> TileIterator<'_, T>
Iterate over all tiles.
§cuda-tile-behavior.md References
- Falsification test #37: Tile iteration covers all elements
Sourcepub fn is_power_of_two_tiles(&self) -> bool
pub fn is_power_of_two_tiles(&self) -> bool
Check if tiles are power-of-two sized.
Power-of-two tiles are preferred for GPU compute.
Sourcepub fn elements_per_tile(&self) -> usize
pub fn elements_per_tile(&self) -> usize
Get the number of elements per tile (maximum).
Sourcepub fn recommended_workgroup_size(&self) -> (u32, u32, u32)
pub fn recommended_workgroup_size(&self) -> (u32, u32, u32)
Get recommended workgroup size for GPU dispatch.
Returns (x, y, z) workgroup dimensions based on tile shape.
Trait Implementations§
Source§impl<T> Clone for PartitionView<T>
impl<T> Clone for PartitionView<T>
Auto Trait Implementations§
impl<T> Freeze for PartitionView<T>
impl<T> RefUnwindSafe for PartitionView<T>where
T: RefUnwindSafe,
impl<T> Send for PartitionView<T>where
T: Send,
impl<T> Sync for PartitionView<T>where
T: Sync,
impl<T> Unpin for PartitionView<T>where
T: Unpin,
impl<T> UnsafeUnpin for PartitionView<T>
impl<T> UnwindSafe for PartitionView<T>where
T: UnwindSafe,
Blanket Implementations§
Source§impl<T> BorrowMut<T> for Twhere
T: ?Sized,
impl<T> BorrowMut<T> for Twhere
T: ?Sized,
Source§fn borrow_mut(&mut self) -> &mut T
fn borrow_mut(&mut self) -> &mut T
Source§impl<T> CloneToUninit for Twhere
T: Clone,
impl<T> CloneToUninit for Twhere
T: Clone,
Source§impl<T> FmtForward for T
impl<T> FmtForward for T
Source§fn fmt_binary(self) -> FmtBinary<Self>where
Self: Binary,
fn fmt_binary(self) -> FmtBinary<Self>where
Self: Binary,
self to use its Binary implementation when Debug-formatted.Source§fn fmt_display(self) -> FmtDisplay<Self>where
Self: Display,
fn fmt_display(self) -> FmtDisplay<Self>where
Self: Display,
self to use its Display implementation when
Debug-formatted.Source§fn fmt_lower_exp(self) -> FmtLowerExp<Self>where
Self: LowerExp,
fn fmt_lower_exp(self) -> FmtLowerExp<Self>where
Self: LowerExp,
self to use its LowerExp implementation when
Debug-formatted.Source§fn fmt_lower_hex(self) -> FmtLowerHex<Self>where
Self: LowerHex,
fn fmt_lower_hex(self) -> FmtLowerHex<Self>where
Self: LowerHex,
self to use its LowerHex implementation when
Debug-formatted.Source§fn fmt_octal(self) -> FmtOctal<Self>where
Self: Octal,
fn fmt_octal(self) -> FmtOctal<Self>where
Self: Octal,
self to use its Octal implementation when Debug-formatted.Source§fn fmt_pointer(self) -> FmtPointer<Self>where
Self: Pointer,
fn fmt_pointer(self) -> FmtPointer<Self>where
Self: Pointer,
self to use its Pointer implementation when
Debug-formatted.Source§fn fmt_upper_exp(self) -> FmtUpperExp<Self>where
Self: UpperExp,
fn fmt_upper_exp(self) -> FmtUpperExp<Self>where
Self: UpperExp,
self to use its UpperExp implementation when
Debug-formatted.Source§fn fmt_upper_hex(self) -> FmtUpperHex<Self>where
Self: UpperHex,
fn fmt_upper_hex(self) -> FmtUpperHex<Self>where
Self: UpperHex,
self to use its UpperHex implementation when
Debug-formatted.Source§impl<T> Instrument for T
impl<T> Instrument for T
Source§fn instrument(self, span: Span) -> Instrumented<Self>
fn instrument(self, span: Span) -> Instrumented<Self>
Source§fn in_current_span(self) -> Instrumented<Self>
fn in_current_span(self) -> Instrumented<Self>
Source§impl<T> IntoEither for T
impl<T> IntoEither for T
Source§fn into_either(self, into_left: bool) -> Either<Self, Self>
fn into_either(self, into_left: bool) -> Either<Self, Self>
self into a Left variant of Either<Self, Self>
if into_left is true.
Converts self into a Right variant of Either<Self, Self>
otherwise. Read moreSource§fn into_either_with<F>(self, into_left: F) -> Either<Self, Self>
fn into_either_with<F>(self, into_left: F) -> Either<Self, Self>
self into a Left variant of Either<Self, Self>
if into_left(&self) returns true.
Converts self into a Right variant of Either<Self, Self>
otherwise. Read moreSource§impl<T> Pipe for Twhere
T: ?Sized,
impl<T> Pipe for Twhere
T: ?Sized,
Source§fn pipe<R>(self, func: impl FnOnce(Self) -> R) -> Rwhere
Self: Sized,
fn pipe<R>(self, func: impl FnOnce(Self) -> R) -> Rwhere
Self: Sized,
Source§fn pipe_ref<'a, R>(&'a self, func: impl FnOnce(&'a Self) -> R) -> Rwhere
R: 'a,
fn pipe_ref<'a, R>(&'a self, func: impl FnOnce(&'a Self) -> R) -> Rwhere
R: 'a,
self and passes that borrow into the pipe function. Read moreSource§fn pipe_ref_mut<'a, R>(&'a mut self, func: impl FnOnce(&'a mut Self) -> R) -> Rwhere
R: 'a,
fn pipe_ref_mut<'a, R>(&'a mut self, func: impl FnOnce(&'a mut Self) -> R) -> Rwhere
R: 'a,
self and passes that borrow into the pipe function. Read moreSource§fn pipe_borrow<'a, B, R>(&'a self, func: impl FnOnce(&'a B) -> R) -> R
fn pipe_borrow<'a, B, R>(&'a self, func: impl FnOnce(&'a B) -> R) -> R
Source§fn pipe_borrow_mut<'a, B, R>(
&'a mut self,
func: impl FnOnce(&'a mut B) -> R,
) -> R
fn pipe_borrow_mut<'a, B, R>( &'a mut self, func: impl FnOnce(&'a mut B) -> R, ) -> R
Source§fn pipe_as_ref<'a, U, R>(&'a self, func: impl FnOnce(&'a U) -> R) -> R
fn pipe_as_ref<'a, U, R>(&'a self, func: impl FnOnce(&'a U) -> R) -> R
self, then passes self.as_ref() into the pipe function.Source§fn pipe_as_mut<'a, U, R>(&'a mut self, func: impl FnOnce(&'a mut U) -> R) -> R
fn pipe_as_mut<'a, U, R>(&'a mut self, func: impl FnOnce(&'a mut U) -> R) -> R
self, then passes self.as_mut() into the pipe
function.Source§fn pipe_deref<'a, T, R>(&'a self, func: impl FnOnce(&'a T) -> R) -> R
fn pipe_deref<'a, T, R>(&'a self, func: impl FnOnce(&'a T) -> R) -> R
self, then passes self.deref() into the pipe function.Source§impl<T> Pointable for T
impl<T> Pointable for T
Source§impl<T> Tap for T
impl<T> Tap for T
Source§fn tap_borrow<B>(self, func: impl FnOnce(&B)) -> Self
fn tap_borrow<B>(self, func: impl FnOnce(&B)) -> Self
Borrow<B> of a value. Read moreSource§fn tap_borrow_mut<B>(self, func: impl FnOnce(&mut B)) -> Self
fn tap_borrow_mut<B>(self, func: impl FnOnce(&mut B)) -> Self
BorrowMut<B> of a value. Read moreSource§fn tap_ref<R>(self, func: impl FnOnce(&R)) -> Self
fn tap_ref<R>(self, func: impl FnOnce(&R)) -> Self
AsRef<R> view of a value. Read moreSource§fn tap_ref_mut<R>(self, func: impl FnOnce(&mut R)) -> Self
fn tap_ref_mut<R>(self, func: impl FnOnce(&mut R)) -> Self
AsMut<R> view of a value. Read moreSource§fn tap_deref<T>(self, func: impl FnOnce(&T)) -> Self
fn tap_deref<T>(self, func: impl FnOnce(&T)) -> Self
Deref::Target of a value. Read moreSource§fn tap_deref_mut<T>(self, func: impl FnOnce(&mut T)) -> Self
fn tap_deref_mut<T>(self, func: impl FnOnce(&mut T)) -> Self
Deref::Target of a value. Read moreSource§fn tap_dbg(self, func: impl FnOnce(&Self)) -> Self
fn tap_dbg(self, func: impl FnOnce(&Self)) -> Self
.tap() only in debug builds, and is erased in release builds.Source§fn tap_mut_dbg(self, func: impl FnOnce(&mut Self)) -> Self
fn tap_mut_dbg(self, func: impl FnOnce(&mut Self)) -> Self
.tap_mut() only in debug builds, and is erased in release
builds.Source§fn tap_borrow_dbg<B>(self, func: impl FnOnce(&B)) -> Self
fn tap_borrow_dbg<B>(self, func: impl FnOnce(&B)) -> Self
.tap_borrow() only in debug builds, and is erased in release
builds.Source§fn tap_borrow_mut_dbg<B>(self, func: impl FnOnce(&mut B)) -> Self
fn tap_borrow_mut_dbg<B>(self, func: impl FnOnce(&mut B)) -> Self
.tap_borrow_mut() only in debug builds, and is erased in release
builds.Source§fn tap_ref_dbg<R>(self, func: impl FnOnce(&R)) -> Self
fn tap_ref_dbg<R>(self, func: impl FnOnce(&R)) -> Self
.tap_ref() only in debug builds, and is erased in release
builds.Source§fn tap_ref_mut_dbg<R>(self, func: impl FnOnce(&mut R)) -> Self
fn tap_ref_mut_dbg<R>(self, func: impl FnOnce(&mut R)) -> Self
.tap_ref_mut() only in debug builds, and is erased in release
builds.Source§fn tap_deref_dbg<T>(self, func: impl FnOnce(&T)) -> Self
fn tap_deref_dbg<T>(self, func: impl FnOnce(&T)) -> Self
.tap_deref() only in debug builds, and is erased in release
builds.