cuDDL
Loading...
Searching...
No Matches
Classes | Namespaces | Macros | Functions
sequence_encode.cuh File Reference
#include <cuda_runtime.h>
#include <cuda/atomic>
#include <cuda/devices>
#include <cuda/functional>
#include <cuda/std/algorithm>
#include <cuda/std/bit>
#include <cuda/std/cstdint>
#include <cuda/stream>
#include <cstdint>
#include <limits>
#include <cub/block/block_reduce.cuh>
#include <cuddl/blacklist.cuh>
#include <cuddl/detail/dna.hpp>
#include <cuddl/detail/hash.cuh>
#include <cuddl/detail/register.cuh>
#include <cuddl/device_span.cuh>
#include <cuddl/error.hpp>

Go to the source code of this file.

Classes

struct  cuddl::detail::sequence_batch_chunk
 One piece of sequence of one genome, and where its bytes are. More...
 

Namespaces

namespace  cuddl
 Batch operations over contiguous device rows.
 
namespace  cuddl::detail
 

Macros

#define CUDDL_SEQUENCE_BATCH_REGISTER_CAP
 

Functions

__device__ __forceinline__ uint32_t cuddl::detail::load_shared_relaxed (uint32_t const *address)
 Relaxed CTA-scope load of a shared word that other threads update atomically.
 
template<size_t BucketCount, typename Layout , bool HasBlacklist = false, bool Naive = false>
__device__ __forceinline__ void cuddl::detail::add_sequence_windows (char const *sequence, uint32_t windows, size_t first_tile, size_t tile_stride, uint32_t k, uint32_t *registers, bool packed_input=false, blacklist_view blacklist={})
 Encodes raw ASCII windows into registers using shared overlapping tiles.
 
template<size_t BucketCount, typename Layout , bool HasBlacklist = false, bool Naive = false>
__global__ void cuddl::detail::add_sequence_batch_kernel (sequence_batch_chunk const *chunks, size_t chunk_count, size_t block_count, uint32_t k, uint32_t *registers, blacklist_view blacklist={})
 Batch kernel covering every staged chunk in one launch, including short records.
 
template<size_t BucketCount, typename Layout , bool HasBlacklist = false>
__global__ void cuddl::detail::add_sequence_single_kernel (char const *sequence, uint32_t windows, uint32_t k, uint32_t *registers, blacklist_view blacklist={})
 Single-sequence kernel for the public raw-ASCII API.
 
template<size_t BucketCount, typename Layout = default_register_layout>
__host__ Result< void > cuddl::detail::launch_sequence_add (device_span< char const > sequence, uint32_t k, device_span< uint32_t > registers, cuda::stream_ref stream, cuda::std::optional< blacklist_view > blacklist=cuda::std::nullopt)
 Shared launch for one device-resident ASCII chunk.
 

Macro Definition Documentation

◆ CUDDL_SEQUENCE_BATCH_REGISTER_CAP

#define CUDDL_SEQUENCE_BATCH_REGISTER_CAP

Definition at line 296 of file sequence_encode.cuh.