 
     basic_data_space< int_type > *  dSubPop;
     pairwise_diff_stats         *  dStats;
+
+    dim3    m_threads, m_blocks;
 };
 
 void upload_sequences( device_sequence_space< int_type > & seqs, device_sequence_space< int_type > *dSeqs ) {

    resize_space( dSeqs, seqs.seq_width, seqs.seq_count );
    upload_sequences( seqs, dSeqs );

    test_kernel t1;

    t1( dSeqs, seqs.seq_count );

    pairwise_diff_stats exp_stats;
    compute_expectation( seqs, exp_stats );

    t1.validate( exp_stats );

    if( seqs.sequences ) {
        delete seqs.sequences;

}

struct test_kernel {
    test_kernel( unsigned int tx = 32, unsigned int ty = 32) :
        m_threads(tx,ty,1)
        , m_blocks( 10, 1, 1)
    {
        create_space( dSubPop );
        assert( cudaMalloc( (void **)&dStats, sizeof(pairwise_diff_stats) ) == cudaSuccess);
    }
