#include <pthread.h>
#include <cstdlib>
#include <iostream>
#include <vector>
#define COMPRESSION_BATCH_SIZE 32

using namespace std;

struct ThreadArg {
  float *original_data;
  long num_elements;
  int thread_num;
  float ***compressed_data;
  bool **compressed_data_taken;
  unsigned int *mask;
};

struct CompressedPos {
  long compressed_data_batch;
  long offset;
};

int n_threads = 8;
int n_compressed_data_batches = 8;

long layer_sizes_alexnet[] = {56l * 56 * 96,  28l * 28 * 96,  27l * 27 * 256,
                              13l * 13 * 256, 13l * 12 * 384, 13l * 12 * 384,
                              13l * 13 * 256, 6l * 6 * 256};
bool layer_compress_alexnet[] = {true, true, true, true,
                                 true, true, true, true};
long layer_density_alexnet[] = {50, 80, 40, 60, 70, 70, 30, 60};
int num_layers_alexnet = 8;

long layer_sizes_vgg[] = {224l * 224 * 64,  224l * 224 * 64,  112l * 112 * 64,
                          112l * 112 * 128, 112l * 112 * 128, 56l * 56 * 128,
                          56l * 56 * 256,   56l * 56 * 256,   56l * 56 * 256,
                          28l * 28 * 256,   28l * 28 * 512,   28l * 28 * 512,
                          28l * 28 * 512,   14l * 14 * 512,   14l * 14 * 512,
                          14l * 14 * 512,   14l * 14 * 512,   7l * 7 * 512};

long layer_density_vgg[] = {50, 20, 30, 20, 10, 20, 20, 20, 10,
                            20, 20, 10, 10, 10, 20, 20, 10, 15};
bool layer_compress_vgg[] = {true, true, true, true, true, true,
                             true, true, true, true, true, true,
                             true, true, true, true, true, true};

int num_layers_vgg = 18;

// long *layer_sizes = layer_sizes_alexnet;
// bool *layer_compress = layer_compress_alexnet;
// long *layer_density = layer_density_alexnet;
// int num_layers = num_layers_alexnet;

long *layer_sizes = layer_sizes_alexnet;
bool *layer_compress = layer_compress_alexnet;
long *layer_density = layer_density_alexnet;
int num_layers = num_layers_alexnet;

void *compressThread(void *arg) {
  ThreadArg *thread_arg = (ThreadArg *)arg;
  float *original_data = thread_arg->original_data;
  float ***compressed_data = thread_arg->compressed_data;
  bool **compressed_data_taken = thread_arg->compressed_data_taken;
  unsigned int *mask = thread_arg->mask;
  int thread_num = thread_arg->thread_num;
  long num_elements = thread_arg->num_elements;

  long start = thread_num * num_elements / n_threads;
  long n_compression_batches =
      num_elements / n_threads / COMPRESSION_BATCH_SIZE;

  long compressed_data_batch_size =
      num_elements / n_threads / n_compressed_data_batches;
  cudaMallocHost((void **)&compressed_data[thread_num],
                 n_compressed_data_batches * sizeof(float *));
  cudaMallocHost((void **)&compressed_data_taken[thread_num],
                 n_compressed_data_batches * sizeof(bool));
  for (int i = 0; i < n_compressed_data_batches; i++) {
    compressed_data_taken[thread_num][i] = false;
  }
  CompressedPos current_pos;
  current_pos.compressed_data_batch = -1,
  current_pos.offset = compressed_data_batch_size;
  for (long i = 0; i < n_compression_batches; i++) {
    long mask_pos =
        (i * COMPRESSION_BATCH_SIZE + start) / COMPRESSION_BATCH_SIZE;
    mask[mask_pos] = 0;
    for (long j = i * COMPRESSION_BATCH_SIZE + start;
         j < (i + 1) * COMPRESSION_BATCH_SIZE + start; j++) {
      if (original_data[j] > 0) {
        if (current_pos.offset == compressed_data_batch_size) {
          cudaMallocHost(
              (void **)&compressed_data[thread_num]
                                       [current_pos.compressed_data_batch + 1],
              compressed_data_batch_size * sizeof(float));
          compressed_data_taken[thread_num]
                               [current_pos.compressed_data_batch + 1] = true;
          current_pos.compressed_data_batch =
              current_pos.compressed_data_batch + 1;
          current_pos.offset = 0;
        }
        mask[mask_pos] = (mask[mask_pos] << 1) + 1;
        compressed_data[thread_num][current_pos.compressed_data_batch]
                       [current_pos.offset] = original_data[j];
        current_pos.offset += 1;
      } else {
        mask[mask_pos] = (mask[mask_pos] << 1);
      }
    }
  }

  return NULL;
}

void *decompressThread(void *arg) {
  ThreadArg *thread_arg = (ThreadArg *)arg;
  float *original_data = thread_arg->original_data;
  float ***compressed_data = thread_arg->compressed_data;
  bool **compressed_data_taken = thread_arg->compressed_data_taken;
  unsigned int *mask = thread_arg->mask;
  int thread_num = thread_arg->thread_num;
  long num_elements = thread_arg->num_elements;

  long start = thread_num * num_elements / n_threads;
  long n_compression_batches =
      num_elements / n_threads / COMPRESSION_BATCH_SIZE;

  long compressed_data_batch_size =
      num_elements / n_threads / n_compressed_data_batches;
  // cudaMallocHost((void **)&compressed_data[thread_num],
  // n_compressed_data_batches * sizeof(float *));
  CompressedPos current_pos;
  current_pos.compressed_data_batch = 0, current_pos.offset = 0;
  for (long i = 0; i < n_compression_batches; i++) {
    long mask_pos =
        (i * COMPRESSION_BATCH_SIZE + start) / COMPRESSION_BATCH_SIZE;
    for (long j = i * COMPRESSION_BATCH_SIZE + start;
         j < (i + 1) * COMPRESSION_BATCH_SIZE + start; j++) {
      if (mask[mask_pos] & 0x80000000 > 0) {
        original_data[j] =
            compressed_data[thread_num][current_pos.compressed_data_batch]
                           [current_pos.offset];
        current_pos.offset += 1;
        if (current_pos.offset == compressed_data_batch_size) {
          current_pos.compressed_data_batch += 1;
          current_pos.offset = 0;
        }
      } else {
        original_data[j] = 0;
      }
      mask[mask_pos] = mask[mask_pos] << 1;
    }
  }
  for (int i = 0; i < n_compressed_data_batches; i++) {
    if (compressed_data_taken[thread_num][i])
      cudaFreeHost(compressed_data[thread_num][i]);
    else
      break;
  }
  cudaFreeHost(compressed_data_taken[thread_num]);
  cudaFreeHost(compressed_data[thread_num]);
  return NULL;
}

int main() {
  int batch_size = 64;
  long total_space = 0;
  cudaEvent_t start, stop;
  cudaEventCreate(&start);
  cudaEventCreate(&stop);
  vector<float ***> compressed_data_vec;
  vector<unsigned int *> mask_vec;
  vector<bool **> compressed_data_taken_vec;

  pthread_t threads[n_threads];
  for (int i = 0; i < num_layers; i++) {
    layer_sizes[i] *= batch_size;
  }
  vector<float> compression_times;
  float total_milli = 0.0;
  for (int j = 0; j < num_layers; j++) {
    if (!layer_compress[j]) continue;
    long num_elements = layer_sizes[j];
    float *original_data, ***compressed_data;
    bool **compressed_data_taken;
    unsigned int *mask;
    cudaMallocHost((void **)&original_data, num_elements * sizeof(float));
    // cudaMallocHost((void **)&compressed_data, num_elements * sizeof(float));

    // generate data
    for (long i = 0; i < num_elements; i++) {
      if (rand() % 100 < layer_density[j])
        original_data[i] = 1;
      else
        original_data[i] = 0;
    }
    if (num_elements % n_threads != 0) {
      cout << "bad number of threads" << endl;
      exit(0);
    }
    if ((num_elements / n_threads) % COMPRESSION_BATCH_SIZE != 0) {
      cout << "bad num_elements or n_threads" << endl;
      exit(0);
    }
    cout << "starting " << j << endl;
    cudaEventRecord(start);

    cudaMallocHost((void **)&compressed_data, n_threads * sizeof(float **));
    cudaMallocHost((void **)&mask, num_elements / COMPRESSION_BATCH_SIZE *
                                       sizeof(unsigned int));
    cudaMallocHost((void **)&compressed_data_taken, n_threads * sizeof(bool *));
    ThreadArg thread_arg[n_threads];
    for (int i = 0; i < n_threads; i++) {
      thread_arg[i].original_data = original_data;
      thread_arg[i].compressed_data = compressed_data;
      thread_arg[i].compressed_data_taken = compressed_data_taken;
      thread_arg[i].mask = mask;
      thread_arg[i].thread_num = i;
      thread_arg[i].num_elements = num_elements;
    }

    for (int i = 0; i < n_threads; i++) {
      pthread_create(&threads[i], NULL, &compressThread,
                     (void *)&thread_arg[i]);
    }
    for (int i = 0; i < n_threads; i++) {
      pthread_join(threads[i], NULL);
    }
    compressed_data_vec.push_back(compressed_data);
    mask_vec.push_back(mask);
    compressed_data_taken_vec.push_back(compressed_data_taken);
    cudaFreeHost(original_data);
    // for (int i = 0; i < 27 * 27 * 256 * 128; i++);
    cudaEventRecord(stop);
    cudaEventSynchronize(stop);
    float milli;
    cudaEventElapsedTime(&milli, start, stop);
    compression_times.push_back(milli);
    total_milli += milli;
    // cout << milli << endl;
    cudaFreeHost(original_data);
    // cudaFreeHost(compressed_data);
    // cudaFreeHost(mask);
  }

  for (int i = 0; i < compression_times.size(); i++) {
    cout << compression_times[i] << endl;
  }
  cout << total_milli << endl;

  // calculating space consumed

  int k = 0;
  for (int j = 0; j < num_layers; j++) {
    long num_elements = layer_sizes[j];
    long cur_space = 0;
    if (!layer_compress[j]) {
      cur_space = num_elements * sizeof(float);
      total_space += cur_space;
      continue;
    }
    bool **compressed_data_taken = compressed_data_taken_vec[k];
    long compressed_data_batch_size =
        num_elements / n_threads / n_compressed_data_batches;
    for (int thread_num = 0; thread_num < n_threads; thread_num++) {
      for (int i = 0; i < n_compressed_data_batches; i++) {
        if (compressed_data_taken[thread_num][i])
          cur_space += compressed_data_batch_size;
        else
          break;
      }
    }
    // add size of mask
    cur_space += num_elements / COMPRESSION_BATCH_SIZE;
    cur_space *= sizeof(float);
    total_space += cur_space;
    k++;
  }
  cout << "total_space_compressed(MB): " << total_space * 1.0 / (1024 * 1024)
       << endl;

  // {
  // 	int n;
  // 	cout << "waiting..\n";
  // 	cin >> n;
  // }

  // decompression
  cout << "decompress" << endl;
  vector<float> decompression_times;
  float total_milli_decompress = 0.0;
  for (int j = num_layers - 1; j >= 0; j--) {
    if (!layer_compress[j]) continue;
    long num_elements = layer_sizes[j];
    float *original_data, ***compressed_data;
    bool **compressed_data_taken;
    unsigned int *mask;
    compressed_data = compressed_data_vec.back();
    mask = mask_vec.back();
    compressed_data_taken = compressed_data_taken_vec.back();
    compressed_data_vec.pop_back();
    mask_vec.pop_back();
    compressed_data_taken_vec.pop_back();
    // cudaMallocHost((void **)&compressed_data, num_elements * sizeof(float));

    cout << "starting " << j << endl;
    cudaEventRecord(start);
    cudaMallocHost((void **)&original_data, num_elements * sizeof(float));
    ThreadArg thread_arg[n_threads];
    for (int i = 0; i < n_threads; i++) {
      thread_arg[i].original_data = original_data;
      thread_arg[i].compressed_data = compressed_data;
      thread_arg[i].compressed_data_taken = compressed_data_taken;
      thread_arg[i].mask = mask;
      thread_arg[i].thread_num = i;
      thread_arg[i].num_elements = num_elements;
    }

    for (int i = 0; i < n_threads; i++) {
      pthread_create(&threads[i], NULL, &decompressThread,
                     (void *)&thread_arg[i]);
    }
    for (int i = 0; i < n_threads; i++) {
      pthread_join(threads[i], NULL);
    }
    cudaFreeHost(compressed_data_taken);
    cudaFreeHost(compressed_data);
    cudaFreeHost(mask);
    // cudaFreeHost(original_data);
    // for (int i = 0; i < 27 * 27 * 256 * 128; i++);
    cudaEventRecord(stop);
    cudaEventSynchronize(stop);
    float milli;
    cudaEventElapsedTime(&milli, start, stop);
    decompression_times.insert(decompression_times.begin(), milli);
    total_milli_decompress += milli;
    // cout << milli << endl;
    // cudaFreeHost(compressed_data);
    // cudaFreeHost(mask);
  }

  for (int i = 0; i < decompression_times.size(); i++) {
    cout << decompression_times[i] << endl;
  }
  cout << total_milli_decompress << endl;

  // calculating total space
  total_space = 0;
  for (int j = 0; j < num_layers; j++) {
    long num_elements = layer_sizes[j];
    long cur_space = 0;
    cur_space = num_elements * sizeof(float);
    total_space += cur_space;
  }
  cout << "total space(MB): " << total_space * 1.0 / (1024 * 1024) << endl;
}
	.file	"tmpxft_00390326_00000000-6_compression_time_req_allocate.cudafe1.cpp"
	.text
#APP
	.globl _ZSt21ios_base_library_initv

#NO_APP
	.globl	_Z14compressThreadPv
	.type	_Z14compressThreadPv, @function
_Z14compressThreadPv:
.LFB3998:
	.cfi_startproc
	endbr64
	pushq	%r15
	.cfi_def_cfa_offset 16
	.cfi_offset 15, -16
	pushq	%r14
	.cfi_def_cfa_offset 24
	.cfi_offset 14, -24
	pushq	%r13
	.cfi_def_cfa_offset 32
	.cfi_offset 13, -32
	pushq	%r12
	.cfi_def_cfa_offset 40
	.cfi_offset 12, -40
	pushq	%rbp
	.cfi_def_cfa_offset 48
	.cfi_offset 6, -48
	pushq	%rbx
	.cfi_def_cfa_offset 56
	.cfi_offset 3, -56
	subq	$72, %rsp
	.cfi_def_cfa_offset 128
	movq	40(%rdi), %rax
	movslq	16(%rdi), %rbp
	movq	8(%rdi), %rbx
	movslq	n_threads(%rip), %rsi
	movq	24(%rdi), %r14
	movq	32(%rdi), %r13
	movq	%rax, 8(%rsp)
	movq	%rbp, %rax
	imulq	%rbx, %rax
	salq	$3, %rbp
	movq	(%rdi), %r15
	addq	%rbp, %r14
	addq	%rbp, %r13
	movq	%r14, %rdi
	cqto
	idivq	%rsi
	movq	%rax, %r12
	movq	%rbx, %rax
	cqto
	idivq	%rsi
	movl	$32, %esi
	cqto
	movq	%rax, %rbx
	idivq	%rsi
	movslq	n_compressed_data_batches(%rip), %rsi
	movq	%rax, 16(%rsp)
	movq	%rbx, %rax
	cqto
	idivq	%rsi
	salq	$3, %rsi
	movq	%rax, %rbx
	call	cudaMallocHost@PLT
	movslq	n_compressed_data_batches(%rip), %rsi
	movq	%r13, %rdi
	call	cudaMallocHost@PLT
	movl	n_compressed_data_batches(%rip), %edx
	xorl	%eax, %eax
.L2:
	cmpl	%eax, %edx
	jle	.L16
	movq	0(%r13), %rcx
	movb	$0, (%rcx,%rax)
	incq	%rax
	jmp	.L2
.L16:
	leaq	0(,%rbx,4), %rax
	movq	%rbx, %rsi
	orq	$-1, %rcx
	xorl	%r8d, %r8d
	movq	%rax, 24(%rsp)
.L4:
	movq	16(%rsp), %rax
	cmpq	%rax, %r8
	jge	.L17
	movq	%r12, %rax
	movl	$32, %edi
	movq	%r12, %rbp
	incq	%r8
	cqto
	leaq	32(%r12), %r12
	idivq	%rdi
	movq	8(%rsp), %rdi
	leaq	(%rdi,%rax,4), %rdx
	xorl	%eax, %eax
	movl	%eax, (%rdx)
.L5:
	cmpq	%r12, %rbp
	jge	.L4
	movss	(%r15,%rbp,4), %xmm1
	xorps	%xmm0, %xmm0
	comiss	%xmm0, %xmm1
	jbe	.L13
	cmpq	%rsi, %rbx
	jne	.L8
	movq	(%r14), %rax
	leaq	1(%rcx), %r11
	movq	24(%rsp), %rsi
	movq	%r8, 56(%rsp)
	movq	%rdx, 48(%rsp)
	leaq	(%rax,%r11,8), %rdi
	movq	%rcx, 40(%rsp)
	movq	%r11, 32(%rsp)
	call	cudaMallocHost@PLT
	movq	40(%rsp), %rcx
	movq	0(%r13), %rax
	xorl	%esi, %esi
	movq	56(%rsp), %r8
	movq	48(%rsp), %rdx
	movb	$1, 1(%rax,%rcx)
	movq	32(%rsp), %rcx
.L8:
	movl	(%rdx), %eax
	movss	(%r15,%rbp,4), %xmm0
	leal	1(%rax,%rax), %eax
	movl	%eax, (%rdx)
	movq	(%r14), %rax
	movq	(%rax,%rcx,8), %rax
	movss	%xmm0, (%rax,%rsi,4)
	incq	%rsi
	jmp	.L9
.L13:
	movd	(%rdx), %xmm0
	pslld	$1, %xmm0
	movd	%xmm0, (%rdx)
.L9:
	incq	%rbp
	jmp	.L5
.L17:
	addq	$72, %rsp
	.cfi_def_cfa_offset 56
	xorl	%eax, %eax
	popq	%rbx
	.cfi_def_cfa_offset 48
	popq	%rbp
	.cfi_def_cfa_offset 40
	popq	%r12
	.cfi_def_cfa_offset 32
	popq	%r13
	.cfi_def_cfa_offset 24
	popq	%r14
	.cfi_def_cfa_offset 16
	popq	%r15
	.cfi_def_cfa_offset 8
	ret
	.cfi_endproc
.LFE3998:
	.size	_Z14compressThreadPv, .-_Z14compressThreadPv
	.globl	_Z16decompressThreadPv
	.type	_Z16decompressThreadPv, @function
_Z16decompressThreadPv:
.LFB3999:
	.cfi_startproc
	endbr64
	pushq	%r15
	.cfi_def_cfa_offset 16
	.cfi_offset 15, -16
	movq	%rdi, %rcx
	movl	$32, %r9d
	movl	$32, %r11d
	pushq	%r14
	.cfi_def_cfa_offset 24
	.cfi_offset 14, -24
	pushq	%r13
	.cfi_def_cfa_offset 32
	.cfi_offset 13, -32
	pushq	%r12
	.cfi_def_cfa_offset 40
	.cfi_offset 12, -40
	pushq	%rbp
	.cfi_def_cfa_offset 48
	.cfi_offset 6, -48
	pushq	%rbx
	.cfi_def_cfa_offset 56
	.cfi_offset 3, -56
	subq	$24, %rsp
	.cfi_def_cfa_offset 80
	movq	8(%rdi), %r8
	movq	(%rdi), %r12
	movq	32(%rdi), %rbp
	movq	40(%rdi), %r13
	movq	%r8, %rax
	movslq	n_threads(%rip), %rdi
	movq	24(%rcx), %rbx
	cqto
	idivq	%rdi
	cqto
	movq	%rax, %rsi
	idivq	%r9
	movslq	n_compressed_data_batches(%rip), %r9
	movq	%rax, 8(%rsp)
	movq	%rsi, %rax
	cqto
	idivq	%r9
	xorl	%r9d, %r9d
	movq	%rax, %rsi
	movslq	16(%rcx), %rax
	leaq	0(,%rax,8), %r10
	imulq	%r8, %rax
	addq	%r10, %rbx
	xorl	%r8d, %r8d
	cqto
	idivq	%rdi
	xorl	%edi, %edi
.L19:
	movq	8(%rsp), %rdx
	cmpq	%rdx, %r8
	jge	.L30
	cqto
	movq	%rax, %rcx
	incq	%r8
	idivq	%r11
	leaq	32(%rcx), %rdx
	leaq	0(%r13,%rax,4), %r14
.L20:
	cmpq	%rcx, %rdx
	jle	.L31
	movl	(%r14), %eax
	xorps	%xmm0, %xmm0
	testb	$1, %al
	je	.L21
	movq	(%rbx), %r15
	movq	(%r15,%r9,8), %r15
	movss	(%r15,%rdi,4), %xmm0
	incq	%rdi
	cmpq	%rsi, %rdi
	jne	.L21
	incq	%r9
	xorl	%edi, %edi
.L21:
	addl	%eax, %eax
	movss	%xmm0, (%r12,%rcx,4)
	incq	%rcx
	movl	%eax, (%r14)
	jmp	.L20
.L31:
	movq	%rdx, %rax
	jmp	.L19
.L30:
	leaq	0(%rbp,%r10), %r12
	xorl	%ebp, %ebp
.L24:
	cmpl	%ebp, n_compressed_data_batches(%rip)
	movq	(%r12), %rdi
	jle	.L25
	cmpb	$0, (%rdi,%rbp)
	je	.L25
	movq	(%rbx), %rax
	movq	(%rax,%rbp,8), %rdi
	incq	%rbp
	call	cudaFreeHost@PLT
	jmp	.L24
.L25:
	call	cudaFreeHost@PLT
	movq	(%rbx), %rdi
	call	cudaFreeHost@PLT
	addq	$24, %rsp
	.cfi_def_cfa_offset 56
	xorl	%eax, %eax
	popq	%rbx
	.cfi_def_cfa_offset 48
	popq	%rbp
	.cfi_def_cfa_offset 40
	popq	%r12
	.cfi_def_cfa_offset 32
	popq	%r13
	.cfi_def_cfa_offset 24
	popq	%r14
	.cfi_def_cfa_offset 16
	popq	%r15
	.cfi_def_cfa_offset 8
	ret
	.cfi_endproc
.LFE3999:
	.size	_Z16decompressThreadPv, .-_Z16decompressThreadPv
	.type	_ZL26__cudaUnregisterBinaryUtilv, @function
_ZL26__cudaUnregisterBinaryUtilv:
.LFB4043:
	.cfi_startproc
	endbr64
	movq	_ZL20__cudaFatCubinHandle(%rip), %rdi
	jmp	__cudaUnregisterFatBinary@PLT
	.cfi_endproc
.LFE4043:
	.size	_ZL26__cudaUnregisterBinaryUtilv, .-_ZL26__cudaUnregisterBinaryUtilv
	.align 2
	.type	_ZNSt12_Vector_baseIPPPfSaIS2_EE11_M_allocateEm.part.0.isra.0, @function
_ZNSt12_Vector_baseIPPPfSaIS2_EE11_M_allocateEm.part.0.isra.0:
.LFB5024:
	.cfi_startproc
	movabsq	$1152921504606846975, %rax
	cmpq	%rdi, %rax
	jnb	.L34
	pushq	%rax
	.cfi_def_cfa_offset 16
	movabsq	$2305843009213693951, %rax
	cmpq	%rdi, %rax
	jnb	.L35
	call	_ZSt28__throw_bad_array_new_lengthv@PLT
.L35:
	call	_ZSt17__throw_bad_allocv@PLT
.L34:
	.cfi_def_cfa_offset 8
	salq	$3, %rdi
	jmp	_Znwm@PLT
	.cfi_endproc
.LFE5024:
	.size	_ZNSt12_Vector_baseIPPPfSaIS2_EE11_M_allocateEm.part.0.isra.0, .-_ZNSt12_Vector_baseIPPPfSaIS2_EE11_M_allocateEm.part.0.isra.0
	.align 2
	.type	_ZNSolsEPFRSoS_E.isra.0, @function
_ZNSolsEPFRSoS_E.isra.0:
.LFB5026:
	.cfi_startproc
	jmp	*%rsi
	.cfi_endproc
.LFE5026:
	.size	_ZNSolsEPFRSoS_E.isra.0, .-_ZNSolsEPFRSoS_E.isra.0
	.section	.text.startup,"ax",@progbits
	.type	_ZL24__sti____cudaRegisterAllv, @function
_ZL24__sti____cudaRegisterAllv:
.LFB4066:
	.cfi_startproc
	endbr64
	pushq	%rax
	.cfi_def_cfa_offset 16
	leaq	_ZL15__fatDeviceText(%rip), %rdi
	call	__cudaRegisterFatBinary@PLT
	movq	%rax, %rdi
	movq	%rax, _ZL20__cudaFatCubinHandle(%rip)
	call	__cudaRegisterFatBinaryEnd@PLT
	leaq	_ZL26__cudaUnregisterBinaryUtilv(%rip), %rdi
	popq	%rdx
	.cfi_def_cfa_offset 8
	jmp	atexit@PLT
	.cfi_endproc
.LFE4066:
	.size	_ZL24__sti____cudaRegisterAllv, .-_ZL24__sti____cudaRegisterAllv
	.section	.init_array,"aw"
	.align 8
	.quad	_ZL24__sti____cudaRegisterAllv
	.section	.text._ZNSt6vectorIPPPfSaIS2_EED2Ev,"axG",@progbits,_ZNSt6vectorIPPPfSaIS2_EED5Ev,comdat
	.align 2
	.weak	_ZNSt6vectorIPPPfSaIS2_EED2Ev
	.type	_ZNSt6vectorIPPPfSaIS2_EED2Ev, @function
_ZNSt6vectorIPPPfSaIS2_EED2Ev:
.LFB4376:
	.cfi_startproc
	endbr64
	movq	(%rdi), %rax
	testq	%rax, %rax
	je	.L42
	movq	16(%rdi), %rsi
	movq	%rax, %rdi
	subq	%rax, %rsi
	jmp	_ZdlPvm@PLT
.L42:
	ret
	.cfi_endproc
.LFE4376:
	.size	_ZNSt6vectorIPPPfSaIS2_EED2Ev, .-_ZNSt6vectorIPPPfSaIS2_EED2Ev
	.weak	_ZNSt6vectorIPPPfSaIS2_EED1Ev
	.set	_ZNSt6vectorIPPPfSaIS2_EED1Ev,_ZNSt6vectorIPPPfSaIS2_EED2Ev
	.section	.text._ZNSt6vectorIPjSaIS0_EED2Ev,"axG",@progbits,_ZNSt6vectorIPjSaIS0_EED5Ev,comdat
	.align 2
	.weak	_ZNSt6vectorIPjSaIS0_EED2Ev
	.type	_ZNSt6vectorIPjSaIS0_EED2Ev, @function
_ZNSt6vectorIPjSaIS0_EED2Ev:
.LFB4388:
	.cfi_startproc
	endbr64
	movq	(%rdi), %rax
	testq	%rax, %rax
	je	.L44
	movq	16(%rdi), %rsi
	movq	%rax, %rdi
	subq	%rax, %rsi
	jmp	_ZdlPvm@PLT
.L44:
	ret
	.cfi_endproc
.LFE4388:
	.size	_ZNSt6vectorIPjSaIS0_EED2Ev, .-_ZNSt6vectorIPjSaIS0_EED2Ev
	.weak	_ZNSt6vectorIPjSaIS0_EED1Ev
	.set	_ZNSt6vectorIPjSaIS0_EED1Ev,_ZNSt6vectorIPjSaIS0_EED2Ev
	.section	.text._ZNSt6vectorIPPbSaIS1_EED2Ev,"axG",@progbits,_ZNSt6vectorIPPbSaIS1_EED5Ev,comdat
	.align 2
	.weak	_ZNSt6vectorIPPbSaIS1_EED2Ev
	.type	_ZNSt6vectorIPPbSaIS1_EED2Ev, @function
_ZNSt6vectorIPPbSaIS1_EED2Ev:
.LFB4400:
	.cfi_startproc
	endbr64
	movq	(%rdi), %rax
	testq	%rax, %rax
	je	.L46
	movq	16(%rdi), %rsi
	movq	%rax, %rdi
	subq	%rax, %rsi
	jmp	_ZdlPvm@PLT
.L46:
	ret
	.cfi_endproc
.LFE4400:
	.size	_ZNSt6vectorIPPbSaIS1_EED2Ev, .-_ZNSt6vectorIPPbSaIS1_EED2Ev
	.weak	_ZNSt6vectorIPPbSaIS1_EED1Ev
	.set	_ZNSt6vectorIPPbSaIS1_EED1Ev,_ZNSt6vectorIPPbSaIS1_EED2Ev
	.section	.text._ZNSt6vectorIfSaIfEED2Ev,"axG",@progbits,_ZNSt6vectorIfSaIfEED5Ev,comdat
	.align 2
	.weak	_ZNSt6vectorIfSaIfEED2Ev
	.type	_ZNSt6vectorIfSaIfEED2Ev, @function
_ZNSt6vectorIfSaIfEED2Ev:
.LFB4412:
	.cfi_startproc
	endbr64
	movq	(%rdi), %rax
	testq	%rax, %rax
	je	.L48
	movq	16(%rdi), %rsi
	movq	%rax, %rdi
	subq	%rax, %rsi
	jmp	_ZdlPvm@PLT
.L48:
	ret
	.cfi_endproc
.LFE4412:
	.size	_ZNSt6vectorIfSaIfEED2Ev, .-_ZNSt6vectorIfSaIfEED2Ev
	.weak	_ZNSt6vectorIfSaIfEED1Ev
	.set	_ZNSt6vectorIfSaIfEED1Ev,_ZNSt6vectorIfSaIfEED2Ev
	.section	.rodata._ZNSt6vectorIPPPfSaIS2_EE17_M_realloc_insertIJRKS2_EEEvN9__gnu_cxx17__normal_iteratorIPS2_S4_EEDpOT_.str1.1,"aMS",@progbits,1
.LC1:
	.string	"vector::_M_realloc_insert"
	.section	.text._ZNSt6vectorIPPPfSaIS2_EE17_M_realloc_insertIJRKS2_EEEvN9__gnu_cxx17__normal_iteratorIPS2_S4_EEDpOT_,"axG",@progbits,_ZNSt6vectorIPPPfSaIS2_EE17_M_realloc_insertIJRKS2_EEEvN9__gnu_cxx17__normal_iteratorIPS2_S4_EEDpOT_,comdat
	.align 2
	.weak	_ZNSt6vectorIPPPfSaIS2_EE17_M_realloc_insertIJRKS2_EEEvN9__gnu_cxx17__normal_iteratorIPS2_S4_EEDpOT_
	.type	_ZNSt6vectorIPPPfSaIS2_EE17_M_realloc_insertIJRKS2_EEEvN9__gnu_cxx17__normal_iteratorIPS2_S4_EEDpOT_, @function
_ZNSt6vectorIPPPfSaIS2_EE17_M_realloc_insertIJRKS2_EEEvN9__gnu_cxx17__normal_iteratorIPS2_S4_EEDpOT_:
.LFB4633:
	.cfi_startproc
	endbr64
	movabsq	$1152921504606846975, %rcx
	pushq	%r15
	.cfi_def_cfa_offset 16
	.cfi_offset 15, -16
	pushq	%r14
	.cfi_def_cfa_offset 24
	.cfi_offset 14, -24
	pushq	%r13
	.cfi_def_cfa_offset 32
	.cfi_offset 13, -32
	pushq	%r12
	.cfi_def_cfa_offset 40
	.cfi_offset 12, -40
	pushq	%rbp
	.cfi_def_cfa_offset 48
	.cfi_offset 6, -48
	pushq	%rbx
	.cfi_def_cfa_offset 56
	.cfi_offset 3, -56
	subq	$24, %rsp
	.cfi_def_cfa_offset 80
	movq	8(%rdi), %rax
	movq	(%rdi), %r14
	movq	%rax, %rbx
	movq	%rax, (%rsp)
	subq	%r14, %rbx
	sarq	$3, %rbx
	cmpq	%rcx, %rbx
	jne	.L51
	leaq	.LC1(%rip), %rdi
	call	_ZSt20__throw_length_errorPKc@PLT
.L51:
	cmpq	%r14, (%rsp)
	movl	$1, %eax
	movq	%rsi, %r15
	movq	%rdi, %r13
	cmovne	%rbx, %rax
	movq	%rsi, %rbp
	addq	%rax, %rbx
	setc	%al
	subq	%r14, %r15
	movzbl	%al, %eax
	testq	%rax, %rax
	jne	.L61
	xorl	%r12d, %r12d
	testq	%rbx, %rbx
	je	.L56
	movabsq	$1152921504606846975, %rax
	cmpq	%rax, %rbx
	cmova	%rax, %rbx
	jmp	.L55
.L61:
	movq	%rcx, %rbx
.L55:
	movq	%rbx, %rdi
	movq	%rdx, 8(%rsp)
	call	_ZNSt12_Vector_baseIPPPfSaIS2_EE11_M_allocateEm.part.0.isra.0
	movq	8(%rsp), %rdx
	movq	%rax, %r12
.L56:
	movq	(%rdx), %rax
	movq	%rax, (%r12,%r15)
	testq	%r15, %r15
	jle	.L57
	movq	%r15, %rdx
	movq	%r14, %rsi
	movq	%r12, %rdi
	call	memmove@PLT
.L57:
	movq	(%rsp), %rax
	leaq	8(%r12,%r15), %rdx
	subq	%rbp, %rax
	testq	%rax, %rax
	jle	.L58
	movq	%rdx, %rdi
	movq	%rbp, %rsi
	movq	%rax, %rcx
	rep movsb
.L58:
	leaq	(%rdx,%rax), %rbp
	testq	%r14, %r14
	je	.L59
	movq	16(%r13), %rsi
	movq	%r14, %rdi
	subq	%r14, %rsi
	call	_ZdlPvm@PLT
.L59:
	leaq	(%r12,%rbx,8), %rax
	movq	%r12, 0(%r13)
	movq	%rbp, 8(%r13)
	movq	%rax, 16(%r13)
	addq	$24, %rsp
	.cfi_def_cfa_offset 56
	popq	%rbx
	.cfi_def_cfa_offset 48
	popq	%rbp
	.cfi_def_cfa_offset 40
	popq	%r12
	.cfi_def_cfa_offset 32
	popq	%r13
	.cfi_def_cfa_offset 24
	popq	%r14
	.cfi_def_cfa_offset 16
	popq	%r15
	.cfi_def_cfa_offset 8
	ret
	.cfi_endproc
.LFE4633:
	.size	_ZNSt6vectorIPPPfSaIS2_EE17_M_realloc_insertIJRKS2_EEEvN9__gnu_cxx17__normal_iteratorIPS2_S4_EEDpOT_, .-_ZNSt6vectorIPPPfSaIS2_EE17_M_realloc_insertIJRKS2_EEEvN9__gnu_cxx17__normal_iteratorIPS2_S4_EEDpOT_
	.section	.text._ZNSt6vectorIPjSaIS0_EE17_M_realloc_insertIJRKS0_EEEvN9__gnu_cxx17__normal_iteratorIPS0_S2_EEDpOT_,"axG",@progbits,_ZNSt6vectorIPjSaIS0_EE17_M_realloc_insertIJRKS0_EEEvN9__gnu_cxx17__normal_iteratorIPS0_S2_EEDpOT_,comdat
	.align 2
	.weak	_ZNSt6vectorIPjSaIS0_EE17_M_realloc_insertIJRKS0_EEEvN9__gnu_cxx17__normal_iteratorIPS0_S2_EEDpOT_
	.type	_ZNSt6vectorIPjSaIS0_EE17_M_realloc_insertIJRKS0_EEEvN9__gnu_cxx17__normal_iteratorIPS0_S2_EEDpOT_, @function
_ZNSt6vectorIPjSaIS0_EE17_M_realloc_insertIJRKS0_EEEvN9__gnu_cxx17__normal_iteratorIPS0_S2_EEDpOT_:
.LFB4639:
	.cfi_startproc
	endbr64
	movabsq	$1152921504606846975, %rcx
	pushq	%r15
	.cfi_def_cfa_offset 16
	.cfi_offset 15, -16
	pushq	%r14
	.cfi_def_cfa_offset 24
	.cfi_offset 14, -24
	pushq	%r13
	.cfi_def_cfa_offset 32
	.cfi_offset 13, -32
	pushq	%r12
	.cfi_def_cfa_offset 40
	.cfi_offset 12, -40
	pushq	%rbp
	.cfi_def_cfa_offset 48
	.cfi_offset 6, -48
	pushq	%rbx
	.cfi_def_cfa_offset 56
	.cfi_offset 3, -56
	subq	$24, %rsp
	.cfi_def_cfa_offset 80
	movq	8(%rdi), %rax
	movq	(%rdi), %r14
	movq	%rax, %rbx
	movq	%rax, (%rsp)
	subq	%r14, %rbx
	sarq	$3, %rbx
	cmpq	%rcx, %rbx
	jne	.L69
	leaq	.LC1(%rip), %rdi
	call	_ZSt20__throw_length_errorPKc@PLT
.L69:
	cmpq	%r14, (%rsp)
	movl	$1, %eax
	movq	%rsi, %r15
	movq	%rdi, %r13
	cmovne	%rbx, %rax
	movq	%rsi, %rbp
	addq	%rax, %rbx
	setc	%al
	subq	%r14, %r15
	movzbl	%al, %eax
	testq	%rax, %rax
	jne	.L79
	xorl	%r12d, %r12d
	testq	%rbx, %rbx
	je	.L74
	movabsq	$1152921504606846975, %rax
	cmpq	%rax, %rbx
	cmova	%rax, %rbx
	jmp	.L73
.L79:
	movq	%rcx, %rbx
.L73:
	movq	%rbx, %rdi
	movq	%rdx, 8(%rsp)
	call	_ZNSt12_Vector_baseIPPPfSaIS2_EE11_M_allocateEm.part.0.isra.0
	movq	8(%rsp), %rdx
	movq	%rax, %r12
.L74:
	movq	(%rdx), %rax
	movq	%rax, (%r12,%r15)
	testq	%r15, %r15
	jle	.L75
	movq	%r15, %rdx
	movq	%r14, %rsi
	movq	%r12, %rdi
	call	memmove@PLT
.L75:
	movq	(%rsp), %rax
	leaq	8(%r12,%r15), %rdx
	subq	%rbp, %rax
	testq	%rax, %rax
	jle	.L76
	movq	%rdx, %rdi
	movq	%rbp, %rsi
	movq	%rax, %rcx
	rep movsb
.L76:
	leaq	(%rdx,%rax), %rbp
	testq	%r14, %r14
	je	.L77
	movq	16(%r13), %rsi
	movq	%r14, %rdi
	subq	%r14, %rsi
	call	_ZdlPvm@PLT
.L77:
	leaq	(%r12,%rbx,8), %rax
	movq	%r12, 0(%r13)
	movq	%rbp, 8(%r13)
	movq	%rax, 16(%r13)
	addq	$24, %rsp
	.cfi_def_cfa_offset 56
	popq	%rbx
	.cfi_def_cfa_offset 48
	popq	%rbp
	.cfi_def_cfa_offset 40
	popq	%r12
	.cfi_def_cfa_offset 32
	popq	%r13
	.cfi_def_cfa_offset 24
	popq	%r14
	.cfi_def_cfa_offset 16
	popq	%r15
	.cfi_def_cfa_offset 8
	ret
	.cfi_endproc
.LFE4639:
	.size	_ZNSt6vectorIPjSaIS0_EE17_M_realloc_insertIJRKS0_EEEvN9__gnu_cxx17__normal_iteratorIPS0_S2_EEDpOT_, .-_ZNSt6vectorIPjSaIS0_EE17_M_realloc_insertIJRKS0_EEEvN9__gnu_cxx17__normal_iteratorIPS0_S2_EEDpOT_
	.section	.text._ZNSt6vectorIPPbSaIS1_EE17_M_realloc_insertIJRKS1_EEEvN9__gnu_cxx17__normal_iteratorIPS1_S3_EEDpOT_,"axG",@progbits,_ZNSt6vectorIPPbSaIS1_EE17_M_realloc_insertIJRKS1_EEEvN9__gnu_cxx17__normal_iteratorIPS1_S3_EEDpOT_,comdat
	.align 2
	.weak	_ZNSt6vectorIPPbSaIS1_EE17_M_realloc_insertIJRKS1_EEEvN9__gnu_cxx17__normal_iteratorIPS1_S3_EEDpOT_
	.type	_ZNSt6vectorIPPbSaIS1_EE17_M_realloc_insertIJRKS1_EEEvN9__gnu_cxx17__normal_iteratorIPS1_S3_EEDpOT_, @function
_ZNSt6vectorIPPbSaIS1_EE17_M_realloc_insertIJRKS1_EEEvN9__gnu_cxx17__normal_iteratorIPS1_S3_EEDpOT_:
.LFB4645:
	.cfi_startproc
	endbr64
	movabsq	$1152921504606846975, %rcx
	pushq	%r15
	.cfi_def_cfa_offset 16
	.cfi_offset 15, -16
	pushq	%r14
	.cfi_def_cfa_offset 24
	.cfi_offset 14, -24
	pushq	%r13
	.cfi_def_cfa_offset 32
	.cfi_offset 13, -32
	pushq	%r12
	.cfi_def_cfa_offset 40
	.cfi_offset 12, -40
	pushq	%rbp
	.cfi_def_cfa_offset 48
	.cfi_offset 6, -48
	pushq	%rbx
	.cfi_def_cfa_offset 56
	.cfi_offset 3, -56
	subq	$24, %rsp
	.cfi_def_cfa_offset 80
	movq	8(%rdi), %rax
	movq	(%rdi), %r14
	movq	%rax, %rbx
	movq	%rax, (%rsp)
	subq	%r14, %rbx
	sarq	$3, %rbx
	cmpq	%rcx, %rbx
	jne	.L87
	leaq	.LC1(%rip), %rdi
	call	_ZSt20__throw_length_errorPKc@PLT
.L87:
	cmpq	%r14, (%rsp)
	movl	$1, %eax
	movq	%rsi, %r15
	movq	%rdi, %r13
	cmovne	%rbx, %rax
	movq	%rsi, %rbp
	addq	%rax, %rbx
	setc	%al
	subq	%r14, %r15
	movzbl	%al, %eax
	testq	%rax, %rax
	jne	.L97
	xorl	%r12d, %r12d
	testq	%rbx, %rbx
	je	.L92
	movabsq	$1152921504606846975, %rax
	cmpq	%rax, %rbx
	cmova	%rax, %rbx
	jmp	.L91
.L97:
	movq	%rcx, %rbx
.L91:
	movq	%rbx, %rdi
	movq	%rdx, 8(%rsp)
	call	_ZNSt12_Vector_baseIPPPfSaIS2_EE11_M_allocateEm.part.0.isra.0
	movq	8(%rsp), %rdx
	movq	%rax, %r12
.L92:
	movq	(%rdx), %rax
	movq	%rax, (%r12,%r15)
	testq	%r15, %r15
	jle	.L93
	movq	%r15, %rdx
	movq	%r14, %rsi
	movq	%r12, %rdi
	call	memmove@PLT
.L93:
	movq	(%rsp), %rax
	leaq	8(%r12,%r15), %rdx
	subq	%rbp, %rax
	testq	%rax, %rax
	jle	.L94
	movq	%rdx, %rdi
	movq	%rbp, %rsi
	movq	%rax, %rcx
	rep movsb
.L94:
	leaq	(%rdx,%rax), %rbp
	testq	%r14, %r14
	je	.L95
	movq	16(%r13), %rsi
	movq	%r14, %rdi
	subq	%r14, %rsi
	call	_ZdlPvm@PLT
.L95:
	leaq	(%r12,%rbx,8), %rax
	movq	%r12, 0(%r13)
	movq	%rbp, 8(%r13)
	movq	%rax, 16(%r13)
	addq	$24, %rsp
	.cfi_def_cfa_offset 56
	popq	%rbx
	.cfi_def_cfa_offset 48
	popq	%rbp
	.cfi_def_cfa_offset 40
	popq	%r12
	.cfi_def_cfa_offset 32
	popq	%r13
	.cfi_def_cfa_offset 24
	popq	%r14
	.cfi_def_cfa_offset 16
	popq	%r15
	.cfi_def_cfa_offset 8
	ret
	.cfi_endproc
.LFE4645:
	.size	_ZNSt6vectorIPPbSaIS1_EE17_M_realloc_insertIJRKS1_EEEvN9__gnu_cxx17__normal_iteratorIPS1_S3_EEDpOT_, .-_ZNSt6vectorIPPbSaIS1_EE17_M_realloc_insertIJRKS1_EEEvN9__gnu_cxx17__normal_iteratorIPS1_S3_EEDpOT_
	.section	.text._ZNSt6vectorIfSaIfEE17_M_realloc_insertIJRKfEEEvN9__gnu_cxx17__normal_iteratorIPfS1_EEDpOT_,"axG",@progbits,_ZNSt6vectorIfSaIfEE17_M_realloc_insertIJRKfEEEvN9__gnu_cxx17__normal_iteratorIPfS1_EEDpOT_,comdat
	.align 2
	.weak	_ZNSt6vectorIfSaIfEE17_M_realloc_insertIJRKfEEEvN9__gnu_cxx17__normal_iteratorIPfS1_EEDpOT_
	.type	_ZNSt6vectorIfSaIfEE17_M_realloc_insertIJRKfEEEvN9__gnu_cxx17__normal_iteratorIPfS1_EEDpOT_, @function
_ZNSt6vectorIfSaIfEE17_M_realloc_insertIJRKfEEEvN9__gnu_cxx17__normal_iteratorIPfS1_EEDpOT_:
.LFB4651:
	.cfi_startproc
	endbr64
	movabsq	$2305843009213693951, %rcx
	pushq	%r15
	.cfi_def_cfa_offset 16
	.cfi_offset 15, -16
	pushq	%r14
	.cfi_def_cfa_offset 24
	.cfi_offset 14, -24
	pushq	%r13
	.cfi_def_cfa_offset 32
	.cfi_offset 13, -32
	pushq	%r12
	.cfi_def_cfa_offset 40
	.cfi_offset 12, -40
	pushq	%rbp
	.cfi_def_cfa_offset 48
	.cfi_offset 6, -48
	pushq	%rbx
	.cfi_def_cfa_offset 56
	.cfi_offset 3, -56
	subq	$24, %rsp
	.cfi_def_cfa_offset 80
	movq	8(%rdi), %rax
	movq	(%rdi), %r14
	movq	%rax, %rbx
	movq	%rax, (%rsp)
	subq	%r14, %rbx
	sarq	$2, %rbx
	cmpq	%rcx, %rbx
	jne	.L105
	leaq	.LC1(%rip), %rdi
	call	_ZSt20__throw_length_errorPKc@PLT
.L105:
	cmpq	%r14, (%rsp)
	movl	$1, %eax
	movq	%rsi, %r15
	movq	%rdi, %r13
	cmovne	%rbx, %rax
	movq	%rsi, %rbp
	addq	%rax, %rbx
	setc	%al
	subq	%r14, %r15
	movzbl	%al, %eax
	testq	%rax, %rax
	jne	.L115
	xorl	%r12d, %r12d
	testq	%rbx, %rbx
	je	.L110
	movabsq	$2305843009213693951, %rax
	cmpq	%rax, %rbx
	cmova	%rax, %rbx
	jmp	.L109
.L115:
	movq	%rcx, %rbx
.L109:
	leaq	0(,%rbx,4), %rdi
	movq	%rdx, 8(%rsp)
	call	_Znwm@PLT
	movq	8(%rsp), %rdx
	movq	%rax, %r12
.L110:
	movss	(%rdx), %xmm0
	movss	%xmm0, (%r12,%r15)
	testq	%r15, %r15
	jle	.L111
	movq	%r15, %rdx
	movq	%r14, %rsi
	movq	%r12, %rdi
	call	memmove@PLT
.L111:
	movq	(%rsp), %rax
	leaq	4(%r12,%r15), %rdx
	subq	%rbp, %rax
	testq	%rax, %rax
	jle	.L112
	movq	%rdx, %rdi
	movq	%rbp, %rsi
	movq	%rax, %rcx
	rep movsb
.L112:
	leaq	(%rdx,%rax), %rbp
	testq	%r14, %r14
	je	.L113
	movq	16(%r13), %rsi
	movq	%r14, %rdi
	subq	%r14, %rsi
	call	_ZdlPvm@PLT
.L113:
	leaq	(%r12,%rbx,4), %rax
	movq	%r12, 0(%r13)
	movq	%rbp, 8(%r13)
	movq	%rax, 16(%r13)
	addq	$24, %rsp
	.cfi_def_cfa_offset 56
	popq	%rbx
	.cfi_def_cfa_offset 48
	popq	%rbp
	.cfi_def_cfa_offset 40
	popq	%r12
	.cfi_def_cfa_offset 32
	popq	%r13
	.cfi_def_cfa_offset 24
	popq	%r14
	.cfi_def_cfa_offset 16
	popq	%r15
	.cfi_def_cfa_offset 8
	ret
	.cfi_endproc
.LFE4651:
	.size	_ZNSt6vectorIfSaIfEE17_M_realloc_insertIJRKfEEEvN9__gnu_cxx17__normal_iteratorIPfS1_EEDpOT_, .-_ZNSt6vectorIfSaIfEE17_M_realloc_insertIJRKfEEEvN9__gnu_cxx17__normal_iteratorIPfS1_EEDpOT_
	.section	.rodata.str1.1,"aMS",@progbits,1
.LC3:
	.string	"bad number of threads"
.LC4:
	.string	"bad num_elements or n_threads"
.LC5:
	.string	"starting "
.LC6:
	.string	"total_space_compressed(MB): "
.LC8:
	.string	"decompress"
.LC9:
	.string	"total space(MB): "
	.section	.text.startup
	.globl	main
	.type	main, @function
main:
.LFB4000:
	.cfi_startproc
	.cfi_personality 0x9b,DW.ref.__gxx_personality_v0
	.cfi_lsda 0x1b,.LLSDA4000
	endbr64
	pushq	%rbp
	.cfi_def_cfa_offset 16
	.cfi_offset 6, -16
	movq	%rsp, %rbp
	.cfi_def_cfa_register 6
	pushq	%r15
	pushq	%r14
	leaq	-216(%rbp), %rdi
	pushq	%r13
	pushq	%r12
	pushq	%rbx
	subq	$232, %rsp
	.cfi_offset 15, -24
	.cfi_offset 14, -32
	.cfi_offset 13, -40
	.cfi_offset 12, -48
	.cfi_offset 3, -56
	movq	%fs:40, %rax
	movq	%rax, -56(%rbp)
	xorl	%eax, %eax
.LEHB0:
	call	cudaEventCreate@PLT
	leaq	-208(%rbp), %rdi
	call	cudaEventCreate@PLT
.LEHE0:
	movslq	n_threads(%rip), %rax
	xorl	%ecx, %ecx
	movq	%rcx, -176(%rbp)
	leaq	15(,%rax,8), %rdx
	movq	%rcx, -168(%rbp)
	movq	%rdx, %rax
	movq	%rcx, -160(%rbp)
	andq	$-4096, %rdx
	movq	%rcx, -152(%rbp)
	andq	$-16, %rax
	movq	%rcx, -144(%rbp)
	movq	%rcx, -136(%rbp)
	movq	%rcx, -128(%rbp)
	movq	%rcx, -120(%rbp)
	movq	%rcx, -112(%rbp)
	movq	%rsp, %rcx
	subq	%rdx, %rcx
.L123:
	cmpq	%rcx, %rsp
	je	.L124
	subq	$4096, %rsp
	orq	$0, 4088(%rsp)
	jmp	.L123
.L124:
	andl	$4095, %eax
	subq	%rax, %rsp
	testq	%rax, %rax
	je	.L125
	orq	$0, -8(%rsp,%rax)
.L125:
	movq	%rsp, -232(%rbp)
	movl	num_layers(%rip), %ecx
	xorl	%eax, %eax
	movq	layer_sizes(%rip), %rdx
.L126:
	cmpl	%eax, %ecx
	jle	.L219
	movq	(%rdx,%rax,8), %xmm0
	psllq	$6, %xmm0
	movq	%xmm0, (%rdx,%rax,8)
	incq	%rax
	jmp	.L126
.L219:
	movl	$0x00000000, -240(%rbp)
	xorl	%edx, %edx
	xorl	%ebx, %ebx
	movq	%rdx, -104(%rbp)
	movq	%rdx, -96(%rbp)
	movq	%rdx, -88(%rbp)
.L128:
	cmpl	%ebx, num_layers(%rip)
	jle	.L220
	movq	layer_compress(%rip), %rax
	movq	%rsp, -248(%rbp)
	cmpb	$0, (%rax,%rbx)
	je	.L130
	movq	layer_sizes(%rip), %rax
	leaq	-200(%rbp), %rdi
	leaq	0(,%rbx,8), %r13
	movq	(%rax,%rbx,8), %r12
	leaq	0(,%r12,4), %rsi
.LEHB1:
	call	cudaMallocHost@PLT
	xorl	%r14d, %r14d
.L131:
	cmpq	%r12, %r14
	jge	.L221
	call	rand@PLT
	movl	$100, %ecx
	cltd
	idivl	%ecx
	movq	layer_density(%rip), %rax
	movslq	%edx, %rdx
	cmpq	(%rax,%r13), %rdx
	movq	-200(%rbp), %rax
	jge	.L132
	movl	$0x3f800000, (%rax,%r14,4)
	jmp	.L133
.L132:
	movl	$0x00000000, (%rax,%r14,4)
.L133:
	incq	%r14
	jmp	.L131
.L221:
	movq	%r12, %rax
	movslq	n_threads(%rip), %rcx
	leaq	_ZSt4cout(%rip), %rdi
	cqto
	idivq	%rcx
	testq	%rdx, %rdx
	je	.L135
	leaq	.LC3(%rip), %rsi
.L218:
	call	_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc@PLT
	movq	_ZSt4endlIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_@GOTPCREL(%rip), %rsi
	movq	%rax, %rdi
	call	_ZNSolsEPFRSoS_E.isra.0
	xorl	%edi, %edi
	call	exit@PLT
.L135:
	testb	$31, %al
	je	.L136
	leaq	.LC4(%rip), %rsi
	jmp	.L218
.L136:
	leaq	.LC5(%rip), %rsi
	call	_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc@PLT
	movq	%rax, %rdi
	movl	%ebx, %esi
	call	_ZNSolsEi@PLT
	movq	_ZSt4endlIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_@GOTPCREL(%rip), %rsi
	movq	%rax, %rdi
	call	_ZNSolsEPFRSoS_E.isra.0
	movq	-216(%rbp), %rdi
	xorl	%esi, %esi
	call	cudaEventRecord@PLT
	movslq	n_threads(%rip), %rsi
	leaq	-192(%rbp), %r15
	movq	%r15, %rdi
	salq	$3, %rsi
	call	cudaMallocHost@PLT
	movq	%r12, %rax
	movl	$32, %ecx
	cqto
	idivq	%rcx
	leaq	0(,%rax,4), %rsi
	leaq	-80(%rbp), %rax
	movq	%rax, %rdi
	movq	%rax, -256(%rbp)
	call	cudaMallocHost@PLT
	movslq	n_threads(%rip), %rsi
	leaq	-184(%rbp), %rdi
	salq	$3, %rsi
	call	cudaMallocHost@PLT
	movslq	n_threads(%rip), %rax
	movq	%rsp, %rdx
	movq	%rax, %rcx
	imulq	$48, %rax, %rax
	movq	%rax, %rsi
	andq	$-4096, %rsi
	subq	%rsi, %rdx
.L138:
	cmpq	%rdx, %rsp
	je	.L139
	subq	$4096, %rsp
	orq	$0, 4088(%rsp)
	jmp	.L138
.L139:
	andl	$4095, %eax
	subq	%rax, %rsp
	testq	%rax, %rax
	je	.L140
	orq	$0, -8(%rsp,%rax)
.L140:
	movq	%rsp, %r13
	movq	-200(%rbp), %rsi
	movq	-80(%rbp), %r9
	xorl	%edx, %edx
	movq	-192(%rbp), %rdi
	movq	-184(%rbp), %r8
	movq	%r13, %rax
.L141:
	cmpl	%ecx, %edx
	jge	.L222
	movl	%edx, 16(%rax)
	incl	%edx
	addq	$48, %rax
	movq	%rsi, -48(%rax)
	movq	%rdi, -24(%rax)
	movq	%r8, -16(%rax)
	movq	%r9, -8(%rax)
	movq	%r12, -40(%rax)
	jmp	.L141
.L222:
	movq	-232(%rbp), %r14
	xorl	%r12d, %r12d
.L143:
	cmpl	%r12d, n_threads(%rip)
	jle	.L223
	movq	%r13, %rcx
	movq	%r14, %rdi
	leaq	_Z14compressThreadPv(%rip), %rdx
	xorl	%esi, %esi
	call	pthread_create@PLT
	incl	%r12d
	addq	$8, %r14
	addq	$48, %r13
	jmp	.L143
.L223:
	xorl	%r12d, %r12d
.L145:
	cmpl	%r12d, n_threads(%rip)
	jle	.L224
	movq	-232(%rbp), %rax
	xorl	%esi, %esi
	movq	(%rax,%r12,8), %rdi
	call	pthread_join@PLT
	incq	%r12
	jmp	.L145
.L224:
	movq	-168(%rbp), %rsi
	cmpq	-160(%rbp), %rsi
	je	.L147
	movq	-192(%rbp), %rax
	addq	$8, %rsi
	movq	%rax, -8(%rsi)
	movq	%rsi, -168(%rbp)
	jmp	.L148
.L147:
	leaq	-176(%rbp), %rdi
	movq	%r15, %rdx
	call	_ZNSt6vectorIPPPfSaIS2_EE17_M_realloc_insertIJRKS2_EEEvN9__gnu_cxx17__normal_iteratorIPS2_S4_EEDpOT_
.L148:
	movq	-144(%rbp), %rsi
	cmpq	-136(%rbp), %rsi
	je	.L149
	movq	-80(%rbp), %rax
	addq	$8, %rsi
	movq	%rax, -8(%rsi)
	movq	%rsi, -144(%rbp)
	jmp	.L150
.L149:
	movq	-256(%rbp), %rdx
	leaq	-152(%rbp), %rdi
	call	_ZNSt6vectorIPjSaIS0_EE17_M_realloc_insertIJRKS0_EEEvN9__gnu_cxx17__normal_iteratorIPS0_S2_EEDpOT_
.L150:
	movq	-120(%rbp), %rsi
	cmpq	-112(%rbp), %rsi
	je	.L151
	movq	-184(%rbp), %rax
	addq	$8, %rsi
	movq	%rax, -8(%rsi)
	movq	%rsi, -120(%rbp)
	jmp	.L152
.L151:
	leaq	-184(%rbp), %rdx
	leaq	-128(%rbp), %rdi
	call	_ZNSt6vectorIPPbSaIS1_EE17_M_realloc_insertIJRKS1_EEEvN9__gnu_cxx17__normal_iteratorIPS1_S3_EEDpOT_
.L152:
	movq	-200(%rbp), %rdi
	call	cudaFreeHost@PLT
	movq	-208(%rbp), %rdi
	xorl	%esi, %esi
	call	cudaEventRecord@PLT
	movq	-208(%rbp), %rdi
	call	cudaEventSynchronize@PLT
	leaq	-220(%rbp), %r12
	movq	-208(%rbp), %rdx
	movq	-216(%rbp), %rsi
	movq	%r12, %rdi
	call	cudaEventElapsedTime@PLT
	movq	-96(%rbp), %rsi
	cmpq	-88(%rbp), %rsi
	je	.L153
	movss	-220(%rbp), %xmm0
	addq	$4, %rsi
	movss	%xmm0, -4(%rsi)
	movq	%rsi, -96(%rbp)
	jmp	.L154
.L153:
	leaq	-104(%rbp), %rdi
	movq	%r12, %rdx
	call	_ZNSt6vectorIfSaIfEE17_M_realloc_insertIJRKfEEEvN9__gnu_cxx17__normal_iteratorIPfS1_EEDpOT_
.L154:
	movss	-240(%rbp), %xmm2
	addss	-220(%rbp), %xmm2
	movq	-200(%rbp), %rdi
	movss	%xmm2, -240(%rbp)
	call	cudaFreeHost@PLT
.LEHE1:
	movq	-248(%rbp), %rsp
.L130:
	incq	%rbx
	jmp	.L128
.L220:
	xorl	%r12d, %r12d
	leaq	_ZSt4cout(%rip), %rbx
.L156:
	movq	-104(%rbp), %rdx
	movq	-96(%rbp), %rax
	subq	%rdx, %rax
	sarq	$2, %rax
	cmpq	%rax, %r12
	jnb	.L225
	movq	%rbx, %rdi
	cvtss2sd	(%rdx,%r12,4), %xmm0
.LEHB2:
	call	_ZNSo9_M_insertIdEERSoT_@PLT
	movq	_ZSt4endlIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_@GOTPCREL(%rip), %rsi
	movq	%rax, %rdi
	call	_ZNSolsEPFRSoS_E.isra.0
	incq	%r12
	jmp	.L156
.L225:
	movq	%rbx, %rdi
	cvtss2sd	-240(%rbp), %xmm0
	call	_ZNSo9_M_insertIdEERSoT_@PLT
	movq	_ZSt4endlIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_@GOTPCREL(%rip), %rsi
	movq	%rax, %rdi
	call	_ZNSolsEPFRSoS_E.isra.0
	movl	num_layers(%rip), %eax
	movq	-128(%rbp), %r11
	xorl	%esi, %esi
	xorl	%edi, %edi
	movq	layer_compress(%rip), %r15
	xorl	%ebx, %ebx
	movl	%eax, -264(%rbp)
	movq	layer_sizes(%rip), %rax
	movq	%rax, -240(%rbp)
	movslq	n_threads(%rip), %rax
	movq	%rax, -248(%rbp)
	movq	%rax, %r9
	movslq	n_compressed_data_batches(%rip), %rax
	movq	%rax, -256(%rbp)
	movq	%rax, %r10
.L158:
	cmpl	%esi, -264(%rbp)
	jle	.L226
	movq	-240(%rbp), %rax
	cmpb	$0, (%r15,%rsi)
	movq	(%rax,%rsi,8), %rcx
	jne	.L159
	leaq	(%rbx,%rcx,4), %rbx
	jmp	.L160
.L159:
	movslq	%edi, %rax
	xorl	%r8d, %r8d
	movq	(%r11,%rax,8), %r12
	movq	%rcx, %rax
	cqto
	idivq	-248(%rbp)
	cqto
	idivq	-256(%rbp)
	xorl	%edx, %edx
.L161:
	cmpl	%edx, %r9d
	jle	.L227
	xorl	%r13d, %r13d
.L164:
	cmpl	%r13d, %r10d
	jle	.L162
	movq	(%r12,%rdx,8), %r14
	movb	(%r14,%r13), %r14b
	incq	%r13
	testb	%r14b, %r14b
	je	.L162
	addq	%rax, %r8
	jmp	.L164
.L162:
	incq	%rdx
	jmp	.L161
.L227:
	movq	%rcx, %rax
	movl	$32, %ecx
	incl	%edi
	cqto
	idivq	%rcx
	addq	%r8, %rax
	leaq	(%rbx,%rax,4), %rbx
.L160:
	incq	%rsi
	jmp	.L158
.L226:
	leaq	_ZSt4cout(%rip), %r12
	leaq	.LC6(%rip), %rsi
	movq	%r12, %rdi
	call	_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc@PLT
	cvtsi2sdq	%rbx, %xmm0
	mulsd	.LC7(%rip), %xmm0
	movq	%rax, %rdi
	call	_ZNSo9_M_insertIdEERSoT_@PLT
	movq	_ZSt4endlIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_@GOTPCREL(%rip), %rsi
	movq	%rax, %rdi
	call	_ZNSolsEPFRSoS_E.isra.0
	leaq	.LC8(%rip), %rsi
	movq	%r12, %rdi
	call	_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc@PLT
	movq	_ZSt4endlIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_@GOTPCREL(%rip), %rsi
	movq	%rax, %rdi
	call	_ZNSolsEPFRSoS_E.isra.0
.LEHE2:
	movl	$0x00000000, -240(%rbp)
	xorl	%eax, %eax
	movq	%rax, -80(%rbp)
	movq	%rax, -72(%rbp)
	movq	%rax, -64(%rbp)
	movl	num_layers(%rip), %eax
	leal	-1(%rax), %ebx
	movslq	%ebx, %r12
.L166:
	testl	%ebx, %ebx
	js	.L228
	movq	layer_compress(%rip), %rax
	movq	%rsp, -256(%rbp)
	cmpb	$0, (%rax,%r12)
	je	.L168
	movq	-144(%rbp), %rdx
	movq	layer_sizes(%rip), %rax
	leaq	.LC5(%rip), %rsi
	leaq	_ZSt4cout(%rip), %rdi
	movq	-168(%rbp), %rcx
	movq	(%rax,%r12,8), %r13
	movq	-8(%rdx), %rax
	subq	$8, %rdx
	movq	-8(%rcx), %r14
	subq	$8, %rcx
	movq	%rdx, -144(%rbp)
	movq	%rax, -264(%rbp)
	movq	-120(%rbp), %rax
	movq	%rcx, -168(%rbp)
	movq	-8(%rax), %r15
	subq	$8, %rax
	movq	%rax, -120(%rbp)
.LEHB3:
	call	_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc@PLT
	movq	%rax, %rdi
	movl	%ebx, %esi
	call	_ZNSolsEi@PLT
	movq	_ZSt4endlIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_@GOTPCREL(%rip), %rsi
	movq	%rax, %rdi
	call	_ZNSolsEPFRSoS_E.isra.0
	movq	-216(%rbp), %rdi
	xorl	%esi, %esi
	call	cudaEventRecord@PLT
	leaq	0(,%r13,4), %rsi
	leaq	-184(%rbp), %rdi
	call	cudaMallocHost@PLT
	movslq	n_threads(%rip), %rax
	movq	%rsp, %rdx
	movq	%rax, %rcx
	imulq	$48, %rax, %rax
	movq	%rax, %rsi
	andq	$-4096, %rsi
	subq	%rsi, %rdx
.L169:
	cmpq	%rdx, %rsp
	je	.L170
	subq	$4096, %rsp
	orq	$0, 4088(%rsp)
	jmp	.L169
.L170:
	andl	$4095, %eax
	subq	%rax, %rsp
	testq	%rax, %rax
	je	.L171
	orq	$0, -8(%rsp,%rax)
.L171:
	movq	%rsp, -248(%rbp)
	movq	-184(%rbp), %rsi
	movq	%rsp, %rax
	xorl	%edx, %edx
.L172:
	cmpl	%ecx, %edx
	jge	.L229
	movq	-264(%rbp), %rdi
	movl	%edx, 16(%rax)
	incl	%edx
	addq	$48, %rax
	movq	%rsi, -48(%rax)
	movq	%r14, -24(%rax)
	movq	%r15, -16(%rax)
	movq	%rdi, -8(%rax)
	movq	%r13, -40(%rax)
	jmp	.L172
.L229:
	movq	-232(%rbp), %rdi
	xorl	%r13d, %r13d
.L174:
	cmpl	%r13d, n_threads(%rip)
	jle	.L230
	movq	-248(%rbp), %rcx
	leaq	_Z16decompressThreadPv(%rip), %rdx
	xorl	%esi, %esi
	incl	%r13d
	movq	%rdi, -272(%rbp)
	call	pthread_create@PLT
	movq	-272(%rbp), %rdi
	addq	$48, -248(%rbp)
	addq	$8, %rdi
	jmp	.L174
.L230:
	xorl	%r13d, %r13d
.L176:
	cmpl	%r13d, n_threads(%rip)
	jle	.L231
	movq	-232(%rbp), %rax
	xorl	%esi, %esi
	movq	(%rax,%r13,8), %rdi
	call	pthread_join@PLT
	incq	%r13
	jmp	.L176
.L231:
	movq	%r15, %rdi
	call	cudaFreeHost@PLT
	movq	%r14, %rdi
	call	cudaFreeHost@PLT
	movq	-264(%rbp), %rdi
	call	cudaFreeHost@PLT
	movq	-208(%rbp), %rdi
	xorl	%esi, %esi
	call	cudaEventRecord@PLT
	movq	-208(%rbp), %rdi
	call	cudaEventSynchronize@PLT
	leaq	-192(%rbp), %r13
	movq	-208(%rbp), %rdx
	movq	-216(%rbp), %rsi
	movq	%r13, %rdi
	call	cudaEventElapsedTime@PLT
	movq	-80(%rbp), %r14
	movq	-72(%rbp), %rdi
	cmpq	-64(%rbp), %rdi
	je	.L178
	movss	-192(%rbp), %xmm0
	leaq	4(%rdi), %rax
	cmpq	%rdi, %r14
	jne	.L179
	movss	%xmm0, (%r14)
	movq	%rax, -72(%rbp)
	jmp	.L180
.L179:
	movss	-4(%rdi), %xmm1
	leaq	-4(%rdi), %rdx
	subq	%r14, %rdx
	movss	%xmm1, (%rdi)
	movq	%rax, -72(%rbp)
	cmpq	$4, %rdx
	jle	.L181
	subq	%rdx, %rdi
	movq	%r14, %rsi
	movss	%xmm0, -248(%rbp)
	call	memmove@PLT
	movss	-248(%rbp), %xmm0
	jmp	.L182
.L181:
	jne	.L182
	movss	(%r14), %xmm1
	movss	%xmm1, -4(%rdi)
.L182:
	movss	%xmm0, (%r14)
	jmp	.L180
.L178:
	leaq	-80(%rbp), %rdi
	movq	%r13, %rdx
	movq	%r14, %rsi
	call	_ZNSt6vectorIfSaIfEE17_M_realloc_insertIJRKfEEEvN9__gnu_cxx17__normal_iteratorIPfS1_EEDpOT_
.LEHE3:
.L180:
	movss	-240(%rbp), %xmm3
	addss	-192(%rbp), %xmm3
	movss	%xmm3, -240(%rbp)
	movq	-256(%rbp), %rsp
.L168:
	decl	%ebx
	decq	%r12
	jmp	.L166
.L228:
	xorl	%r12d, %r12d
	leaq	_ZSt4cout(%rip), %rbx
.L184:
	movq	-80(%rbp), %rdx
	movq	-72(%rbp), %rax
	subq	%rdx, %rax
	sarq	$2, %rax
	cmpq	%rax, %r12
	jnb	.L232
	movq	%rbx, %rdi
	cvtss2sd	(%rdx,%r12,4), %xmm0
.LEHB4:
	call	_ZNSo9_M_insertIdEERSoT_@PLT
	movq	_ZSt4endlIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_@GOTPCREL(%rip), %rsi
	movq	%rax, %rdi
	call	_ZNSolsEPFRSoS_E.isra.0
	incq	%r12
	jmp	.L184
.L232:
	movq	%rbx, %rdi
	cvtss2sd	-240(%rbp), %xmm0
	call	_ZNSo9_M_insertIdEERSoT_@PLT
	movq	_ZSt4endlIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_@GOTPCREL(%rip), %rsi
	movq	%rax, %rdi
	call	_ZNSolsEPFRSoS_E.isra.0
	movl	num_layers(%rip), %edx
	movq	layer_sizes(%rip), %rcx
	xorl	%eax, %eax
	xorl	%r12d, %r12d
.L186:
	cmpl	%eax, %edx
	jle	.L233
	movq	(%rcx,%rax,8), %rsi
	incq	%rax
	leaq	(%r12,%rsi,4), %r12
	jmp	.L186
.L233:
	leaq	.LC9(%rip), %rsi
	movq	%rbx, %rdi
	call	_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc@PLT
	cvtsi2sdq	%r12, %xmm0
	mulsd	.LC7(%rip), %xmm0
	movq	%rax, %rdi
	call	_ZNSo9_M_insertIdEERSoT_@PLT
	movq	_ZSt4endlIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_@GOTPCREL(%rip), %rsi
	movq	%rax, %rdi
	call	_ZNSolsEPFRSoS_E.isra.0
.LEHE4:
	leaq	-80(%rbp), %rdi
	call	_ZNSt6vectorIfSaIfEED1Ev
	leaq	-104(%rbp), %rdi
	call	_ZNSt6vectorIfSaIfEED1Ev
	leaq	-128(%rbp), %rdi
	call	_ZNSt6vectorIPPbSaIS1_EED1Ev
	leaq	-152(%rbp), %rdi
	call	_ZNSt6vectorIPjSaIS0_EED1Ev
	leaq	-176(%rbp), %rdi
	call	_ZNSt6vectorIPPPfSaIS2_EED1Ev
	movq	-56(%rbp), %rax
	subq	%fs:40, %rax
	je	.L193
	jmp	.L213
.L199:
	endbr64
	movq	%rax, %rbx
.L188:
	movq	-248(%rbp), %rsp
	jmp	.L189
.L197:
	endbr64
	movq	%rax, %rbx
.L190:
	movq	-256(%rbp), %rsp
	jmp	.L191
.L198:
	endbr64
	movq	%rax, %rbx
.L191:
	leaq	-80(%rbp), %rdi
	call	_ZNSt6vectorIfSaIfEED1Ev
	jmp	.L189
.L196:
	endbr64
	movq	%rax, %rbx
.L189:
	leaq	-104(%rbp), %rdi
	call	_ZNSt6vectorIfSaIfEED1Ev
	leaq	-128(%rbp), %rdi
	call	_ZNSt6vectorIPPbSaIS1_EED1Ev
	leaq	-152(%rbp), %rdi
	call	_ZNSt6vectorIPjSaIS0_EED1Ev
	leaq	-176(%rbp), %rdi
	call	_ZNSt6vectorIPPPfSaIS2_EED1Ev
	movq	-56(%rbp), %rax
	subq	%fs:40, %rax
	jne	.L213
	movq	%rbx, %rdi
.LEHB5:
	call	_Unwind_Resume@PLT
.LEHE5:
.L213:
	call	__stack_chk_fail@PLT
.L193:
	leaq	-40(%rbp), %rsp
	xorl	%eax, %eax
	popq	%rbx
	popq	%r12
	popq	%r13
	popq	%r14
	popq	%r15
	popq	%rbp
	.cfi_def_cfa 7, 8
	ret
	.cfi_endproc
.LFE4000:
	.globl	__gxx_personality_v0
	.section	.gcc_except_table,"a",@progbits
.LLSDA4000:
	.byte	0xff
	.byte	0xff
	.byte	0x1
	.uleb128 .LLSDACSE4000-.LLSDACSB4000
.LLSDACSB4000:
	.uleb128 .LEHB0-.LFB4000
	.uleb128 .LEHE0-.LEHB0
	.uleb128 0
	.uleb128 0
	.uleb128 .LEHB1-.LFB4000
	.uleb128 .LEHE1-.LEHB1
	.uleb128 .L199-.LFB4000
	.uleb128 0
	.uleb128 .LEHB2-.LFB4000
	.uleb128 .LEHE2-.LEHB2
	.uleb128 .L196-.LFB4000
	.uleb128 0
	.uleb128 .LEHB3-.LFB4000
	.uleb128 .LEHE3-.LEHB3
	.uleb128 .L197-.LFB4000
	.uleb128 0
	.uleb128 .LEHB4-.LFB4000
	.uleb128 .LEHE4-.LEHB4
	.uleb128 .L198-.LFB4000
	.uleb128 0
	.uleb128 .LEHB5-.LFB4000
	.uleb128 .LEHE5-.LEHB5
	.uleb128 0
	.uleb128 0
.LLSDACSE4000:
	.section	.text.startup
	.size	main, .-main
	.type	_GLOBAL__sub_I_n_threads, @function
_GLOBAL__sub_I_n_threads:
.LFB5013:
	.cfi_startproc
	endbr64
	movl	num_layers_alexnet(%rip), %eax
	movl	%eax, num_layers(%rip)
	ret
	.cfi_endproc
.LFE5013:
	.size	_GLOBAL__sub_I_n_threads, .-_GLOBAL__sub_I_n_threads
	.section	.init_array
	.align 8
	.quad	_GLOBAL__sub_I_n_threads
	.section	.nvFatBinSegment,"aw"
	.align 8
	.type	_ZL15__fatDeviceText, @object
	.size	_ZL15__fatDeviceText, 24
_ZL15__fatDeviceText:
	.long	1180844977
	.long	1
	.quad	fatbinData
	.quad	0
	.local	_ZL20__cudaFatCubinHandle
	.comm	_ZL20__cudaFatCubinHandle,8,8
	.globl	num_layers
	.bss
	.align 4
	.type	num_layers, @object
	.size	num_layers, 4
num_layers:
	.zero	4
	.globl	layer_density
	.section	.data.rel.local,"aw"
	.align 8
	.type	layer_density, @object
	.size	layer_density, 8
layer_density:
	.quad	layer_density_alexnet
	.globl	layer_compress
	.align 8
	.type	layer_compress, @object
	.size	layer_compress, 8
layer_compress:
	.quad	layer_compress_alexnet
	.globl	layer_sizes
	.align 8
	.type	layer_sizes, @object
	.size	layer_sizes, 8
layer_sizes:
	.quad	layer_sizes_alexnet
	.globl	num_layers_vgg
	.data
	.align 4
	.type	num_layers_vgg, @object
	.size	num_layers_vgg, 4
num_layers_vgg:
	.long	18
	.globl	layer_compress_vgg
	.align 16
	.type	layer_compress_vgg, @object
	.size	layer_compress_vgg, 18
layer_compress_vgg:
	.byte	1
	.byte	1
	.byte	1
	.byte	1
	.byte	1
	.byte	1
	.byte	1
	.byte	1
	.byte	1
	.byte	1
	.byte	1
	.byte	1
	.byte	1
	.byte	1
	.byte	1
	.byte	1
	.byte	1
	.byte	1
	.globl	layer_density_vgg
	.align 32
	.type	layer_density_vgg, @object
	.size	layer_density_vgg, 144
layer_density_vgg:
	.quad	50
	.quad	20
	.quad	30
	.quad	20
	.quad	10
	.quad	20
	.quad	20
	.quad	20
	.quad	10
	.quad	20
	.quad	20
	.quad	10
	.quad	10
	.quad	10
	.quad	20
	.quad	20
	.quad	10
	.quad	15
	.globl	layer_sizes_vgg
	.align 32
	.type	layer_sizes_vgg, @object
	.size	layer_sizes_vgg, 144
layer_sizes_vgg:
	.quad	3211264
	.quad	3211264
	.quad	802816
	.quad	1605632
	.quad	1605632
	.quad	401408
	.quad	802816
	.quad	802816
	.quad	802816
	.quad	200704
	.quad	401408
	.quad	401408
	.quad	401408
	.quad	100352
	.quad	100352
	.quad	100352
	.quad	100352
	.quad	25088
	.globl	num_layers_alexnet
	.align 4
	.type	num_layers_alexnet, @object
	.size	num_layers_alexnet, 4
num_layers_alexnet:
	.long	8
	.globl	layer_density_alexnet
	.align 32
	.type	layer_density_alexnet, @object
	.size	layer_density_alexnet, 64
layer_density_alexnet:
	.quad	50
	.quad	80
	.quad	40
	.quad	60
	.quad	70
	.quad	70
	.quad	30
	.quad	60
	.globl	layer_compress_alexnet
	.align 8
	.type	layer_compress_alexnet, @object
	.size	layer_compress_alexnet, 8
layer_compress_alexnet:
	.byte	1
	.byte	1
	.byte	1
	.byte	1
	.byte	1
	.byte	1
	.byte	1
	.byte	1
	.globl	layer_sizes_alexnet
	.align 32
	.type	layer_sizes_alexnet, @object
	.size	layer_sizes_alexnet, 64
layer_sizes_alexnet:
	.quad	301056
	.quad	75264
	.quad	186624
	.quad	43264
	.quad	59904
	.quad	59904
	.quad	43264
	.quad	9216
	.globl	n_compressed_data_batches
	.align 4
	.type	n_compressed_data_batches, @object
	.size	n_compressed_data_batches, 4
n_compressed_data_batches:
	.long	8
	.globl	n_threads
	.align 4
	.type	n_threads, @object
	.size	n_threads, 4
n_threads:
	.long	8
	.section	.rodata.cst8,"aM",@progbits,8
	.align 8
.LC7:
	.long	0
	.long	1051721728
	.hidden	DW.ref.__gxx_personality_v0
	.weak	DW.ref.__gxx_personality_v0
	.section	.data.rel.local.DW.ref.__gxx_personality_v0,"awG",@progbits,DW.ref.__gxx_personality_v0,comdat
	.align 8
	.type	DW.ref.__gxx_personality_v0, @object
	.size	DW.ref.__gxx_personality_v0, 8
DW.ref.__gxx_personality_v0:
	.quad	__gxx_personality_v0
	.ident	"GCC: (Ubuntu 13.3.0-6ubuntu2~24.04) 13.3.0"
	.section	.note.GNU-stack,"",@progbits
	.section	.note.gnu.property,"a"
	.align 8
	.long	1f - 0f
	.long	4f - 1f
	.long	5
0:
	.string	"GNU"
1:
	.align 8
	.long	0xc0000002
	.long	3f - 2f
2:
	.long	0x3
3:
	.align 8
4:


#include <hip/hip_runtime.h>
#include <pthread.h>
#include <cstdlib>
#include <iostream>
#include <vector>
#define COMPRESSION_BATCH_SIZE 32

using namespace std;

struct ThreadArg {
  float *original_data;
  long num_elements;
  int thread_num;
  float ***compressed_data;
  bool **compressed_data_taken;
  unsigned int *mask;
};

struct CompressedPos {
  long compressed_data_batch;
  long offset;
};

int n_threads = 8;
int n_compressed_data_batches = 8;

long layer_sizes_alexnet[] = {56l * 56 * 96,  28l * 28 * 96,  27l * 27 * 256,
                              13l * 13 * 256, 13l * 12 * 384, 13l * 12 * 384,
                              13l * 13 * 256, 6l * 6 * 256};
bool layer_compress_alexnet[] = {true, true, true, true,
                                 true, true, true, true};
long layer_density_alexnet[] = {50, 80, 40, 60, 70, 70, 30, 60};
int num_layers_alexnet = 8;

long layer_sizes_vgg[] = {224l * 224 * 64,  224l * 224 * 64,  112l * 112 * 64,
                          112l * 112 * 128, 112l * 112 * 128, 56l * 56 * 128,
                          56l * 56 * 256,   56l * 56 * 256,   56l * 56 * 256,
                          28l * 28 * 256,   28l * 28 * 512,   28l * 28 * 512,
                          28l * 28 * 512,   14l * 14 * 512,   14l * 14 * 512,
                          14l * 14 * 512,   14l * 14 * 512,   7l * 7 * 512};

long layer_density_vgg[] = {50, 20, 30, 20, 10, 20, 20, 20, 10,
                            20, 20, 10, 10, 10, 20, 20, 10, 15};
bool layer_compress_vgg[] = {true, true, true, true, true, true,
                             true, true, true, true, true, true,
                             true, true, true, true, true, true};

int num_layers_vgg = 18;

// long *layer_sizes = layer_sizes_alexnet;
// bool *layer_compress = layer_compress_alexnet;
// long *layer_density = layer_density_alexnet;
// int num_layers = num_layers_alexnet;

long *layer_sizes = layer_sizes_alexnet;
bool *layer_compress = layer_compress_alexnet;
long *layer_density = layer_density_alexnet;
int num_layers = num_layers_alexnet;

void *compressThread(void *arg) {
  ThreadArg *thread_arg = (ThreadArg *)arg;
  float *original_data = thread_arg->original_data;
  float ***compressed_data = thread_arg->compressed_data;
  bool **compressed_data_taken = thread_arg->compressed_data_taken;
  unsigned int *mask = thread_arg->mask;
  int thread_num = thread_arg->thread_num;
  long num_elements = thread_arg->num_elements;

  long start = thread_num * num_elements / n_threads;
  long n_compression_batches =
      num_elements / n_threads / COMPRESSION_BATCH_SIZE;

  long compressed_data_batch_size =
      num_elements / n_threads / n_compressed_data_batches;
  hipHostMalloc((void **)&compressed_data[thread_num],
                 n_compressed_data_batches * sizeof(float *), hipHostMallocDefault);
  hipHostMalloc((void **)&compressed_data_taken[thread_num],
                 n_compressed_data_batches * sizeof(bool), hipHostMallocDefault);
  for (int i = 0; i < n_compressed_data_batches; i++) {
    compressed_data_taken[thread_num][i] = false;
  }
  CompressedPos current_pos;
  current_pos.compressed_data_batch = -1,
  current_pos.offset = compressed_data_batch_size;
  for (long i = 0; i < n_compression_batches; i++) {
    long mask_pos =
        (i * COMPRESSION_BATCH_SIZE + start) / COMPRESSION_BATCH_SIZE;
    mask[mask_pos] = 0;
    for (long j = i * COMPRESSION_BATCH_SIZE + start;
         j < (i + 1) * COMPRESSION_BATCH_SIZE + start; j++) {
      if (original_data[j] > 0) {
        if (current_pos.offset == compressed_data_batch_size) {
          hipHostMalloc(
              (void **)&compressed_data[thread_num]
                                       [current_pos.compressed_data_batch + 1],
              compressed_data_batch_size * sizeof(float), hipHostMallocDefault);
          compressed_data_taken[thread_num]
                               [current_pos.compressed_data_batch + 1] = true;
          current_pos.compressed_data_batch =
              current_pos.compressed_data_batch + 1;
          current_pos.offset = 0;
        }
        mask[mask_pos] = (mask[mask_pos] << 1) + 1;
        compressed_data[thread_num][current_pos.compressed_data_batch]
                       [current_pos.offset] = original_data[j];
        current_pos.offset += 1;
      } else {
        mask[mask_pos] = (mask[mask_pos] << 1);
      }
    }
  }

  return NULL;
}

void *decompressThread(void *arg) {
  ThreadArg *thread_arg = (ThreadArg *)arg;
  float *original_data = thread_arg->original_data;
  float ***compressed_data = thread_arg->compressed_data;
  bool **compressed_data_taken = thread_arg->compressed_data_taken;
  unsigned int *mask = thread_arg->mask;
  int thread_num = thread_arg->thread_num;
  long num_elements = thread_arg->num_elements;

  long start = thread_num * num_elements / n_threads;
  long n_compression_batches =
      num_elements / n_threads / COMPRESSION_BATCH_SIZE;

  long compressed_data_batch_size =
      num_elements / n_threads / n_compressed_data_batches;
  // cudaMallocHost((void **)&compressed_data[thread_num],
  // n_compressed_data_batches * sizeof(float *));
  CompressedPos current_pos;
  current_pos.compressed_data_batch = 0, current_pos.offset = 0;
  for (long i = 0; i < n_compression_batches; i++) {
    long mask_pos =
        (i * COMPRESSION_BATCH_SIZE + start) / COMPRESSION_BATCH_SIZE;
    for (long j = i * COMPRESSION_BATCH_SIZE + start;
         j < (i + 1) * COMPRESSION_BATCH_SIZE + start; j++) {
      if (mask[mask_pos] & 0x80000000 > 0) {
        original_data[j] =
            compressed_data[thread_num][current_pos.compressed_data_batch]
                           [current_pos.offset];
        current_pos.offset += 1;
        if (current_pos.offset == compressed_data_batch_size) {
          current_pos.compressed_data_batch += 1;
          current_pos.offset = 0;
        }
      } else {
        original_data[j] = 0;
      }
      mask[mask_pos] = mask[mask_pos] << 1;
    }
  }
  for (int i = 0; i < n_compressed_data_batches; i++) {
    if (compressed_data_taken[thread_num][i])
      hipHostFree(compressed_data[thread_num][i]);
    else
      break;
  }
  hipHostFree(compressed_data_taken[thread_num]);
  hipHostFree(compressed_data[thread_num]);
  return NULL;
}

int main() {
  int batch_size = 64;
  long total_space = 0;
  hipEvent_t start, stop;
  hipEventCreate(&start);
  hipEventCreate(&stop);
  vector<float ***> compressed_data_vec;
  vector<unsigned int *> mask_vec;
  vector<bool **> compressed_data_taken_vec;

  pthread_t threads[n_threads];
  for (int i = 0; i < num_layers; i++) {
    layer_sizes[i] *= batch_size;
  }
  vector<float> compression_times;
  float total_milli = 0.0;
  for (int j = 0; j < num_layers; j++) {
    if (!layer_compress[j]) continue;
    long num_elements = layer_sizes[j];
    float *original_data, ***compressed_data;
    bool **compressed_data_taken;
    unsigned int *mask;
    hipHostMalloc((void **)&original_data, num_elements * sizeof(float), hipHostMallocDefault);
    // cudaMallocHost((void **)&compressed_data, num_elements * sizeof(float));

    // generate data
    for (long i = 0; i < num_elements; i++) {
      if (rand() % 100 < layer_density[j])
        original_data[i] = 1;
      else
        original_data[i] = 0;
    }
    if (num_elements % n_threads != 0) {
      cout << "bad number of threads" << endl;
      exit(0);
    }
    if ((num_elements / n_threads) % COMPRESSION_BATCH_SIZE != 0) {
      cout << "bad num_elements or n_threads" << endl;
      exit(0);
    }
    cout << "starting " << j << endl;
    hipEventRecord(start);

    hipHostMalloc((void **)&compressed_data, n_threads * sizeof(float **), hipHostMallocDefault);
    hipHostMalloc((void **)&mask, num_elements / COMPRESSION_BATCH_SIZE *
                                       sizeof(unsigned int), hipHostMallocDefault);
    hipHostMalloc((void **)&compressed_data_taken, n_threads * sizeof(bool *), hipHostMallocDefault);
    ThreadArg thread_arg[n_threads];
    for (int i = 0; i < n_threads; i++) {
      thread_arg[i].original_data = original_data;
      thread_arg[i].compressed_data = compressed_data;
      thread_arg[i].compressed_data_taken = compressed_data_taken;
      thread_arg[i].mask = mask;
      thread_arg[i].thread_num = i;
      thread_arg[i].num_elements = num_elements;
    }

    for (int i = 0; i < n_threads; i++) {
      pthread_create(&threads[i], NULL, &compressThread,
                     (void *)&thread_arg[i]);
    }
    for (int i = 0; i < n_threads; i++) {
      pthread_join(threads[i], NULL);
    }
    compressed_data_vec.push_back(compressed_data);
    mask_vec.push_back(mask);
    compressed_data_taken_vec.push_back(compressed_data_taken);
    hipHostFree(original_data);
    // for (int i = 0; i < 27 * 27 * 256 * 128; i++);
    hipEventRecord(stop);
    hipEventSynchronize(stop);
    float milli;
    hipEventElapsedTime(&milli, start, stop);
    compression_times.push_back(milli);
    total_milli += milli;
    // cout << milli << endl;
    hipHostFree(original_data);
    // cudaFreeHost(compressed_data);
    // cudaFreeHost(mask);
  }

  for (int i = 0; i < compression_times.size(); i++) {
    cout << compression_times[i] << endl;
  }
  cout << total_milli << endl;

  // calculating space consumed

  int k = 0;
  for (int j = 0; j < num_layers; j++) {
    long num_elements = layer_sizes[j];
    long cur_space = 0;
    if (!layer_compress[j]) {
      cur_space = num_elements * sizeof(float);
      total_space += cur_space;
      continue;
    }
    bool **compressed_data_taken = compressed_data_taken_vec[k];
    long compressed_data_batch_size =
        num_elements / n_threads / n_compressed_data_batches;
    for (int thread_num = 0; thread_num < n_threads; thread_num++) {
      for (int i = 0; i < n_compressed_data_batches; i++) {
        if (compressed_data_taken[thread_num][i])
          cur_space += compressed_data_batch_size;
        else
          break;
      }
    }
    // add size of mask
    cur_space += num_elements / COMPRESSION_BATCH_SIZE;
    cur_space *= sizeof(float);
    total_space += cur_space;
    k++;
  }
  cout << "total_space_compressed(MB): " << total_space * 1.0 / (1024 * 1024)
       << endl;

  // {
  // 	int n;
  // 	cout << "waiting..\n";
  // 	cin >> n;
  // }

  // decompression
  cout << "decompress" << endl;
  vector<float> decompression_times;
  float total_milli_decompress = 0.0;
  for (int j = num_layers - 1; j >= 0; j--) {
    if (!layer_compress[j]) continue;
    long num_elements = layer_sizes[j];
    float *original_data, ***compressed_data;
    bool **compressed_data_taken;
    unsigned int *mask;
    compressed_data = compressed_data_vec.back();
    mask = mask_vec.back();
    compressed_data_taken = compressed_data_taken_vec.back();
    compressed_data_vec.pop_back();
    mask_vec.pop_back();
    compressed_data_taken_vec.pop_back();
    // cudaMallocHost((void **)&compressed_data, num_elements * sizeof(float));

    cout << "starting " << j << endl;
    hipEventRecord(start);
    hipHostMalloc((void **)&original_data, num_elements * sizeof(float), hipHostMallocDefault);
    ThreadArg thread_arg[n_threads];
    for (int i = 0; i < n_threads; i++) {
      thread_arg[i].original_data = original_data;
      thread_arg[i].compressed_data = compressed_data;
      thread_arg[i].compressed_data_taken = compressed_data_taken;
      thread_arg[i].mask = mask;
      thread_arg[i].thread_num = i;
      thread_arg[i].num_elements = num_elements;
    }

    for (int i = 0; i < n_threads; i++) {
      pthread_create(&threads[i], NULL, &decompressThread,
                     (void *)&thread_arg[i]);
    }
    for (int i = 0; i < n_threads; i++) {
      pthread_join(threads[i], NULL);
    }
    hipHostFree(compressed_data_taken);
    hipHostFree(compressed_data);
    hipHostFree(mask);
    // cudaFreeHost(original_data);
    // for (int i = 0; i < 27 * 27 * 256 * 128; i++);
    hipEventRecord(stop);
    hipEventSynchronize(stop);
    float milli;
    hipEventElapsedTime(&milli, start, stop);
    decompression_times.insert(decompression_times.begin(), milli);
    total_milli_decompress += milli;
    // cout << milli << endl;
    // cudaFreeHost(compressed_data);
    // cudaFreeHost(mask);
  }

  for (int i = 0; i < decompression_times.size(); i++) {
    cout << decompression_times[i] << endl;
  }
  cout << total_milli_decompress << endl;

  // calculating total space
  total_space = 0;
  for (int j = 0; j < num_layers; j++) {
    long num_elements = layer_sizes[j];
    long cur_space = 0;
    cur_space = num_elements * sizeof(float);
    total_space += cur_space;
  }
  cout << "total space(MB): " << total_space * 1.0 / (1024 * 1024) << endl;
}
	.text
	.file	"compression_time_req_allocate.hip"
                                        # Start of file scope inline assembly
	.globl	_ZSt21ios_base_library_initv

                                        # End of file scope inline assembly
	.globl	_Z14compressThreadPv            # -- Begin function _Z14compressThreadPv
	.type	_Z14compressThreadPv,@function
_Z14compressThreadPv:                   # @_Z14compressThreadPv
	.cfi_startproc
# %bb.0:
	pushq	%rbp
	.cfi_def_cfa_offset 16
	pushq	%r15
	.cfi_def_cfa_offset 24
	pushq	%r14
	.cfi_def_cfa_offset 32
	pushq	%r13
	.cfi_def_cfa_offset 40
	pushq	%r12
	.cfi_def_cfa_offset 48
	pushq	%rbx
	.cfi_def_cfa_offset 56
	subq	$72, %rsp
	.cfi_def_cfa_offset 128
	.cfi_offset %rbx, -56
	.cfi_offset %r12, -48
	.cfi_offset %r13, -40
	.cfi_offset %r14, -32
	.cfi_offset %r15, -24
	.cfi_offset %rbp, -16
	movq	(%rdi), %rax
	movq	%rax, 24(%rsp)                  # 8-byte Spill
	movq	8(%rdi), %rcx
	movq	40(%rdi), %r14
	movslq	16(%rdi), %rbx
	movq	%rcx, %rax
	imulq	%rbx, %rax
	movslq	n_threads(%rip), %rsi
	cqto
	idivq	%rsi
	movq	%rax, 40(%rsp)                  # 8-byte Spill
	movq	%rcx, %rax
	cqto
	idivq	%rsi
	movq	%rax, %r13
	addq	$31, %rax
	testq	%r13, %r13
	cmovnsq	%r13, %rax
	movq	%rax, 8(%rsp)                   # 8-byte Spill
	movslq	n_compressed_data_batches(%rip), %rsi
	movq	%r13, %rax
	cqto
	idivq	%rsi
	movq	%rax, 32(%rsp)                  # 8-byte Spill
	shlq	$3, %rbx
	movq	24(%rdi), %rax
	addq	%rbx, %rax
	shlq	$3, %rsi
	addq	32(%rdi), %rbx
	movq	%rax, 16(%rsp)                  # 8-byte Spill
	movq	%rax, %rdi
	xorl	%edx, %edx
	callq	hipHostMalloc
	movslq	n_compressed_data_batches(%rip), %rsi
	movq	%rbx, %rdi
	xorl	%edx, %edx
	callq	hipHostMalloc
	movl	n_compressed_data_batches(%rip), %edx
	testl	%edx, %edx
	jle	.LBB0_2
# %bb.1:                                # %.lr.ph
	movq	(%rbx), %rdi
	xorl	%esi, %esi
	callq	memset@PLT
.LBB0_2:                                # %.preheader
	movq	%rbx, 64(%rsp)                  # 8-byte Spill
	cmpq	$32, %r13
	movq	24(%rsp), %rsi                  # 8-byte Reload
	movq	16(%rsp), %rdi                  # 8-byte Reload
	jl	.LBB0_12
# %bb.3:                                # %.lr.ph71
	sarq	$5, 8(%rsp)                     # 8-byte Folded Spill
	movq	32(%rsp), %rax                  # 8-byte Reload
	leaq	(,%rax,4), %rcx
	movq	%rcx, 56(%rsp)                  # 8-byte Spill
	movq	$-1, %rbp
	xorl	%r12d, %r12d
	xorps	%xmm1, %xmm1
.LBB0_5:                                # %.lr.ph65.preheader
                                        # =>This Loop Header: Depth=1
                                        #     Child Loop BB0_6 Depth 2
	movq	%r12, %r13
	shlq	$5, %r13
	movq	40(%rsp), %rcx                  # 8-byte Reload
	leaq	(%rcx,%r13), %r15
	addq	$31, %r15
	addq	%rcx, %r13
	cmovnsq	%r13, %r15
	sarq	$5, %r15
	movl	$0, (%r14,%r15,4)
	incq	%r12
	movq	%r12, 48(%rsp)                  # 8-byte Spill
	shlq	$5, %r12
	addq	%rcx, %r12
	xorl	%ecx, %ecx
.LBB0_6:                                # %.lr.ph65
                                        #   Parent Loop BB0_5 Depth=1
                                        # =>  This Inner Loop Header: Depth=2
	movss	(%rsi,%r13,4), %xmm0            # xmm0 = mem[0],zero,zero,zero
	ucomiss	%xmm1, %xmm0
	jbe	.LBB0_10
# %bb.7:                                #   in Loop: Header=BB0_6 Depth=2
	movq	(%rdi), %rdx
	cmpq	32(%rsp), %rax                  # 8-byte Folded Reload
	jne	.LBB0_9
# %bb.8:                                #   in Loop: Header=BB0_6 Depth=2
	leaq	1(%rbp), %rbx
	leaq	(%rdx,%rbp,8), %rdi
	addq	$8, %rdi
	movq	56(%rsp), %rsi                  # 8-byte Reload
	xorl	%edx, %edx
	callq	hipHostMalloc
	xorps	%xmm1, %xmm1
	movq	16(%rsp), %rdi                  # 8-byte Reload
	movq	24(%rsp), %rsi                  # 8-byte Reload
	movq	64(%rsp), %rax                  # 8-byte Reload
	movq	(%rax), %rax
	movb	$1, 1(%rax,%rbp)
	movl	(%r14,%r15,4), %ecx
	movss	(%rsi,%r13,4), %xmm0            # xmm0 = mem[0],zero,zero,zero
	movq	(%rdi), %rdx
	xorl	%eax, %eax
	movq	%rbx, %rbp
.LBB0_9:                                #   in Loop: Header=BB0_6 Depth=2
	leal	1(,%rcx,2), %ecx
	movl	%ecx, (%r14,%r15,4)
	movq	(%rdx,%rbp,8), %rdx
	movss	%xmm0, (%rdx,%rax,4)
	incq	%rax
	jmp	.LBB0_11
.LBB0_10:                               #   in Loop: Header=BB0_6 Depth=2
	addl	%ecx, %ecx
	movl	%ecx, (%r14,%r15,4)
.LBB0_11:                               #   in Loop: Header=BB0_6 Depth=2
	incq	%r13
	cmpq	%r12, %r13
	jl	.LBB0_6
# %bb.4:                                # %.loopexit
                                        #   in Loop: Header=BB0_5 Depth=1
	movq	48(%rsp), %r12                  # 8-byte Reload
	cmpq	8(%rsp), %r12                   # 8-byte Folded Reload
	jne	.LBB0_5
.LBB0_12:                               # %._crit_edge
	xorl	%eax, %eax
	addq	$72, %rsp
	.cfi_def_cfa_offset 56
	popq	%rbx
	.cfi_def_cfa_offset 48
	popq	%r12
	.cfi_def_cfa_offset 40
	popq	%r13
	.cfi_def_cfa_offset 32
	popq	%r14
	.cfi_def_cfa_offset 24
	popq	%r15
	.cfi_def_cfa_offset 16
	popq	%rbp
	.cfi_def_cfa_offset 8
	retq
.Lfunc_end0:
	.size	_Z14compressThreadPv, .Lfunc_end0-_Z14compressThreadPv
	.cfi_endproc
                                        # -- End function
	.globl	_Z16decompressThreadPv          # -- Begin function _Z16decompressThreadPv
	.type	_Z16decompressThreadPv,@function
_Z16decompressThreadPv:                 # @_Z16decompressThreadPv
	.cfi_startproc
# %bb.0:
	pushq	%rbp
	.cfi_def_cfa_offset 16
	pushq	%r15
	.cfi_def_cfa_offset 24
	pushq	%r14
	.cfi_def_cfa_offset 32
	pushq	%r13
	.cfi_def_cfa_offset 40
	pushq	%r12
	.cfi_def_cfa_offset 48
	pushq	%rbx
	.cfi_def_cfa_offset 56
	subq	$24, %rsp
	.cfi_def_cfa_offset 80
	.cfi_offset %rbx, -56
	.cfi_offset %r12, -48
	.cfi_offset %r13, -40
	.cfi_offset %r14, -32
	.cfi_offset %r15, -24
	.cfi_offset %rbp, -16
	movq	8(%rdi), %rsi
	movq	24(%rdi), %rbx
	movq	32(%rdi), %r15
	movslq	16(%rdi), %r14
	movq	%rsi, %rax
	imulq	%r14, %rax
	movslq	n_threads(%rip), %r8
	cqto
	idivq	%r8
	movq	%rax, %rcx
	movq	%rsi, %rax
	cqto
	idivq	%r8
	movq	%rax, %rsi
	leaq	31(%rax), %r11
	testq	%rax, %rax
	cmovnsq	%rax, %r11
	movslq	n_compressed_data_batches(%rip), %r9
	cqto
	idivq	%r9
	cmpq	$32, %rsi
	jl	.LBB1_8
# %bb.1:                                # %.lr.ph63
	movq	%r15, 8(%rsp)                   # 8-byte Spill
	movq	(%rdi), %rdx
	movq	40(%rdi), %r8
	sarq	$5, %r11
	xorl	%edi, %edi
	xorl	%r10d, %r10d
	xorl	%r9d, %r9d
	movq	%r11, 16(%rsp)                  # 8-byte Spill
.LBB1_2:                                # %.lr.ph
                                        # =>This Loop Header: Depth=1
                                        #     Child Loop BB1_3 Depth 2
	movq	%r10, %r11
	incq	%r10
	movq	%r10, %r12
	shlq	$5, %r12
	addq	%rcx, %r12
	shlq	$5, %r11
	leaq	(%r11,%rcx), %r13
	addq	$31, %r13
	addq	%rcx, %r11
	cmovnsq	%r11, %r13
	sarq	$5, %r13
	movq	%r8, %rsi
	movl	(%r8,%r13,4), %ebp
.LBB1_3:                                #   Parent Loop BB1_2 Depth=1
                                        # =>  This Inner Loop Header: Depth=2
	xorps	%xmm0, %xmm0
	testb	$1, %bpl
	je	.LBB1_5
# %bb.4:                                #   in Loop: Header=BB1_3 Depth=2
	movq	(%rbx,%r14,8), %r15
	movq	(%r15,%r9,8), %r15
	movss	(%r15,%rdi,4), %xmm0            # xmm0 = mem[0],zero,zero,zero
	incq	%rdi
	xorl	%r15d, %r15d
	cmpq	%rax, %rdi
	sete	%r15b
	movl	$0, %r8d
	cmoveq	%r8, %rdi
	addq	%r15, %r9
.LBB1_5:                                #   in Loop: Header=BB1_3 Depth=2
	movss	%xmm0, (%rdx,%r11,4)
	addl	%ebp, %ebp
	incq	%r11
	cmpq	%r12, %r11
	jl	.LBB1_3
# %bb.6:                                # %.loopexit
                                        #   in Loop: Header=BB1_2 Depth=1
	movl	%ebp, (%rsi,%r13,4)
	movq	%rsi, %r8
	movq	16(%rsp), %r11                  # 8-byte Reload
	cmpq	%r11, %r10
	jne	.LBB1_2
# %bb.7:                                # %.preheader.loopexit
	movl	n_compressed_data_batches(%rip), %r9d
	movq	8(%rsp), %r15                   # 8-byte Reload
.LBB1_8:                                # %.preheader
	testl	%r9d, %r9d
	jle	.LBB1_12
# %bb.9:                                # %.lr.ph65
	xorl	%r12d, %r12d
.LBB1_10:                               # =>This Inner Loop Header: Depth=1
	movq	(%r15,%r14,8), %rax
	cmpb	$0, (%rax,%r12)
	je	.LBB1_12
# %bb.11:                               #   in Loop: Header=BB1_10 Depth=1
	movq	(%rbx,%r14,8), %rax
	movq	(%rax,%r12,8), %rdi
	callq	hipHostFree
	incq	%r12
	movslq	n_compressed_data_batches(%rip), %rax
	cmpq	%rax, %r12
	jl	.LBB1_10
.LBB1_12:                               # %._crit_edge
	movq	(%r15,%r14,8), %rdi
	callq	hipHostFree
	movq	(%rbx,%r14,8), %rdi
	callq	hipHostFree
	xorl	%eax, %eax
	addq	$24, %rsp
	.cfi_def_cfa_offset 56
	popq	%rbx
	.cfi_def_cfa_offset 48
	popq	%r12
	.cfi_def_cfa_offset 40
	popq	%r13
	.cfi_def_cfa_offset 32
	popq	%r14
	.cfi_def_cfa_offset 24
	popq	%r15
	.cfi_def_cfa_offset 16
	popq	%rbp
	.cfi_def_cfa_offset 8
	retq
.Lfunc_end1:
	.size	_Z16decompressThreadPv, .Lfunc_end1-_Z16decompressThreadPv
	.cfi_endproc
                                        # -- End function
	.section	.rodata.cst4,"aM",@progbits,4
	.p2align	2, 0x0                          # -- Begin function main
.LCPI2_0:
	.long	0x3f800000                      # float 1
	.section	.rodata.cst8,"aM",@progbits,8
	.p2align	3, 0x0
.LCPI2_1:
	.quad	0x3eb0000000000000              # double 9.5367431640625E-7
	.text
	.globl	main
	.type	main,@function
main:                                   # @main
.Lfunc_begin0:
	.cfi_startproc
	.cfi_personality 3, __gxx_personality_v0
	.cfi_lsda 3, .Lexception0
# %bb.0:
	pushq	%rbp
	.cfi_def_cfa_offset 16
	.cfi_offset %rbp, -16
	movq	%rsp, %rbp
	.cfi_def_cfa_register %rbp
	pushq	%r15
	pushq	%r14
	pushq	%r13
	pushq	%r12
	pushq	%rbx
	subq	$248, %rsp
	.cfi_offset %rbx, -56
	.cfi_offset %r12, -48
	.cfi_offset %r13, -40
	.cfi_offset %r14, -32
	.cfi_offset %r15, -24
	leaq	-136(%rbp), %rdi
	callq	hipEventCreate
	leaq	-80(%rbp), %rdi
	callq	hipEventCreate
	movq	%rsp, -280(%rbp)                # 8-byte Spill
	xorl	%eax, %eax
	movq	%rax, -256(%rbp)
	xorps	%xmm0, %xmm0
	movaps	%xmm0, -272(%rbp)
	movq	%rax, -224(%rbp)
	movaps	%xmm0, -240(%rbp)
	movq	%rax, -192(%rbp)
	movaps	%xmm0, -208(%rbp)
	movl	n_threads(%rip), %eax
	movq	%rsp, %rcx
	leaq	15(,%rax,8), %rax
	andq	$-16, %rax
	subq	%rax, %rcx
	movq	%rcx, -64(%rbp)                 # 8-byte Spill
	movq	%rcx, %rsp
	movl	num_layers(%rip), %eax
	testl	%eax, %eax
	jle	.LBB2_61
# %bb.1:                                # %.lr.ph
	movq	layer_sizes(%rip), %rcx
	xorl	%edx, %edx
.LBB2_2:                                # =>This Inner Loop Header: Depth=1
	shlq	$6, (%rcx,%rdx,8)
	incq	%rdx
	cmpq	%rdx, %rax
	jne	.LBB2_2
# %bb.3:                                # %._crit_edge
	xorps	%xmm0, %xmm0
	movaps	%xmm0, -160(%rbp)
	movq	$0, -144(%rbp)
	testl	%eax, %eax
	jle	.LBB2_62
# %bb.4:                                # %.lr.ph288
	xorps	%xmm0, %xmm0
	movss	%xmm0, -48(%rbp)                # 4-byte Spill
	movq	layer_compress(%rip), %rcx
	xorl	%ebx, %ebx
.LBB2_5:                                # =>This Loop Header: Depth=1
                                        #     Child Loop BB2_9 Depth 2
                                        #     Child Loop BB2_25 Depth 2
                                        #     Child Loop BB2_28 Depth 2
                                        #     Child Loop BB2_31 Depth 2
	cmpb	$0, (%rcx,%rbx)
	je	.LBB2_52
# %bb.6:                                #   in Loop: Header=BB2_5 Depth=1
	movq	layer_sizes(%rip), %rax
	movq	(%rax,%rbx,8), %r15
	leaq	(,%r15,4), %rsi
.Ltmp0:
	leaq	-128(%rbp), %rdi
	xorl	%edx, %edx
	callq	hipHostMalloc
.Ltmp1:
# %bb.7:                                # %.preheader263
                                        #   in Loop: Header=BB2_5 Depth=1
	testq	%r15, %r15
	jle	.LBB2_12
# %bb.8:                                # %.lr.ph275.preheader
                                        #   in Loop: Header=BB2_5 Depth=1
	xorl	%r12d, %r12d
.LBB2_9:                                # %.lr.ph275
                                        #   Parent Loop BB2_5 Depth=1
                                        # =>  This Inner Loop Header: Depth=2
	callq	rand
	cltq
	imulq	$1374389535, %rax, %rcx         # imm = 0x51EB851F
	movq	%rcx, %rdx
	shrq	$63, %rdx
	sarq	$37, %rcx
	addl	%edx, %ecx
	imull	$100, %ecx, %ecx
	subl	%ecx, %eax
	cltq
	movq	layer_density(%rip), %rcx
	movss	.LCPI2_0(%rip), %xmm0           # xmm0 = mem[0],zero,zero,zero
	cmpq	%rax, (%rcx,%rbx,8)
	jg	.LBB2_11
# %bb.10:                               # %.lr.ph275
                                        #   in Loop: Header=BB2_9 Depth=2
	xorps	%xmm0, %xmm0
.LBB2_11:                               # %.lr.ph275
                                        #   in Loop: Header=BB2_9 Depth=2
	movq	-128(%rbp), %rax
	movss	%xmm0, (%rax,%r12,4)
	incq	%r12
	cmpq	%r12, %r15
	jne	.LBB2_9
.LBB2_12:                               # %._crit_edge276
                                        #   in Loop: Header=BB2_5 Depth=1
	movslq	n_threads(%rip), %rcx
	movq	%r15, %rax
	cqto
	idivq	%rcx
	testq	%rdx, %rdx
	jne	.LBB2_154
# %bb.13:                               #   in Loop: Header=BB2_5 Depth=1
	testb	$31, %al
	jne	.LBB2_152
# %bb.14:                               #   in Loop: Header=BB2_5 Depth=1
.Ltmp11:
	movl	$_ZSt4cout, %edi
	movl	$.L.str.2, %esi
	movl	$9, %edx
	callq	_ZSt16__ostream_insertIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_PKS3_l
.Ltmp12:
# %bb.15:                               # %_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc.exit164
                                        #   in Loop: Header=BB2_5 Depth=1
.Ltmp13:
	movl	$_ZSt4cout, %edi
	movl	%ebx, %esi
	callq	_ZNSolsEi
.Ltmp14:
# %bb.16:                               #   in Loop: Header=BB2_5 Depth=1
	movq	%rax, %r12
	movq	(%rax), %rax
	movq	-24(%rax), %rdi
	addq	%r12, %rdi
.Ltmp15:
	movl	$10, %esi
	callq	_ZNKSt9basic_iosIcSt11char_traitsIcEE5widenEc
.Ltmp16:
# %bb.17:                               # %.noexc221
                                        #   in Loop: Header=BB2_5 Depth=1
.Ltmp17:
	movsbl	%al, %esi
	movq	%r12, %rdi
	callq	_ZNSo3putEc
.Ltmp18:
# %bb.18:                               # %.noexc222
                                        #   in Loop: Header=BB2_5 Depth=1
.Ltmp19:
	movq	%rax, %rdi
	callq	_ZNSo5flushEv
.Ltmp20:
# %bb.19:                               # %_ZNSolsEPFRSoS_E.exit165
                                        #   in Loop: Header=BB2_5 Depth=1
	movq	-136(%rbp), %rdi
.Ltmp21:
	xorl	%esi, %esi
	callq	hipEventRecord
.Ltmp22:
# %bb.20:                               #   in Loop: Header=BB2_5 Depth=1
	movslq	n_threads(%rip), %rsi
	shlq	$3, %rsi
.Ltmp23:
	leaq	-104(%rbp), %rdi
	xorl	%edx, %edx
	callq	hipHostMalloc
.Ltmp24:
# %bb.21:                               #   in Loop: Header=BB2_5 Depth=1
	leaq	31(%r15), %rsi
	testq	%r15, %r15
	cmovnsq	%r15, %rsi
	sarq	$5, %rsi
	shlq	$2, %rsi
.Ltmp25:
	leaq	-176(%rbp), %rdi
	xorl	%edx, %edx
	callq	hipHostMalloc
.Ltmp26:
# %bb.22:                               #   in Loop: Header=BB2_5 Depth=1
	movslq	n_threads(%rip), %rsi
	shlq	$3, %rsi
.Ltmp27:
	leaq	-96(%rbp), %rdi
	xorl	%edx, %edx
	callq	hipHostMalloc
.Ltmp28:
# %bb.23:                               #   in Loop: Header=BB2_5 Depth=1
	movq	%rsp, %r13
	movl	n_threads(%rip), %eax
	movq	%rsp, %r12
	shlq	$4, %rax
	leaq	(%rax,%rax,2), %rax
	subq	%rax, %r12
	movq	%r12, %rsp
	movl	n_threads(%rip), %eax
	testl	%eax, %eax
	jle	.LBB2_33
# %bb.24:                               # %.lr.ph279
                                        #   in Loop: Header=BB2_5 Depth=1
	movq	-128(%rbp), %rcx
	movq	-104(%rbp), %rdx
	movq	-96(%rbp), %rsi
	movq	-176(%rbp), %rdi
	leaq	40(%r12), %r8
	xorl	%r9d, %r9d
.LBB2_25:                               #   Parent Loop BB2_5 Depth=1
                                        # =>  This Inner Loop Header: Depth=2
	movq	%rcx, -40(%r8)
	movq	%rdx, -16(%r8)
	movq	%rsi, -8(%r8)
	movq	%rdi, (%r8)
	movl	%r9d, -24(%r8)
	movq	%r15, -32(%r8)
	incq	%r9
	addq	$48, %r8
	cmpq	%r9, %rax
	jne	.LBB2_25
# %bb.26:                               # %.preheader262
                                        #   in Loop: Header=BB2_5 Depth=1
	testl	%eax, %eax
	jle	.LBB2_33
# %bb.27:                               # %.lr.ph281.preheader
                                        #   in Loop: Header=BB2_5 Depth=1
	movq	-64(%rbp), %r15                 # 8-byte Reload
	xorl	%r14d, %r14d
.LBB2_28:                               # %.lr.ph281
                                        #   Parent Loop BB2_5 Depth=1
                                        # =>  This Inner Loop Header: Depth=2
	movl	$_Z14compressThreadPv, %edx
	movq	%r15, %rdi
	xorl	%esi, %esi
	movq	%r12, %rcx
	callq	pthread_create
	incq	%r14
	movslq	n_threads(%rip), %rax
	addq	$48, %r12
	addq	$8, %r15
	cmpq	%rax, %r14
	jl	.LBB2_28
# %bb.29:                               # %.preheader261
                                        #   in Loop: Header=BB2_5 Depth=1
	testl	%eax, %eax
	jle	.LBB2_33
# %bb.30:                               # %.lr.ph283.preheader
                                        #   in Loop: Header=BB2_5 Depth=1
	xorl	%r14d, %r14d
.LBB2_31:                               # %.lr.ph283
                                        #   Parent Loop BB2_5 Depth=1
                                        # =>  This Inner Loop Header: Depth=2
	movq	-64(%rbp), %rax                 # 8-byte Reload
	movq	(%rax,%r14,8), %rdi
.Ltmp29:
	xorl	%esi, %esi
	callq	pthread_join
.Ltmp30:
# %bb.32:                               #   in Loop: Header=BB2_31 Depth=2
	incq	%r14
	movslq	n_threads(%rip), %rax
	cmpq	%rax, %r14
	jl	.LBB2_31
.LBB2_33:                               # %._crit_edge284
                                        #   in Loop: Header=BB2_5 Depth=1
	movq	-264(%rbp), %rsi
	cmpq	-256(%rbp), %rsi
	je	.LBB2_35
# %bb.34:                               #   in Loop: Header=BB2_5 Depth=1
	movq	-104(%rbp), %rax
	movq	%rax, (%rsi)
	addq	$8, -264(%rbp)
	jmp	.LBB2_36
.LBB2_35:                               #   in Loop: Header=BB2_5 Depth=1
.Ltmp32:
	leaq	-272(%rbp), %rdi
	leaq	-104(%rbp), %rdx
	callq	_ZNSt6vectorIPPPfSaIS2_EE17_M_realloc_insertIJRKS2_EEEvN9__gnu_cxx17__normal_iteratorIPS2_S4_EEDpOT_
.Ltmp33:
.LBB2_36:                               # %_ZNSt6vectorIPPPfSaIS2_EE9push_backERKS2_.exit
                                        #   in Loop: Header=BB2_5 Depth=1
	movq	-232(%rbp), %rsi
	cmpq	-224(%rbp), %rsi
	je	.LBB2_38
# %bb.37:                               #   in Loop: Header=BB2_5 Depth=1
	movq	-176(%rbp), %rax
	movq	%rax, (%rsi)
	addq	$8, -232(%rbp)
	jmp	.LBB2_39
.LBB2_38:                               #   in Loop: Header=BB2_5 Depth=1
.Ltmp34:
	leaq	-240(%rbp), %rdi
	leaq	-176(%rbp), %rdx
	callq	_ZNSt6vectorIPjSaIS0_EE17_M_realloc_insertIJRKS0_EEEvN9__gnu_cxx17__normal_iteratorIPS0_S2_EEDpOT_
.Ltmp35:
.LBB2_39:                               # %_ZNSt6vectorIPjSaIS0_EE9push_backERKS0_.exit
                                        #   in Loop: Header=BB2_5 Depth=1
	movq	-200(%rbp), %rsi
	cmpq	-192(%rbp), %rsi
	je	.LBB2_41
# %bb.40:                               #   in Loop: Header=BB2_5 Depth=1
	movq	-96(%rbp), %rax
	movq	%rax, (%rsi)
	addq	$8, -200(%rbp)
	jmp	.LBB2_42
.LBB2_41:                               #   in Loop: Header=BB2_5 Depth=1
.Ltmp36:
	leaq	-208(%rbp), %rdi
	leaq	-96(%rbp), %rdx
	callq	_ZNSt6vectorIPPbSaIS1_EE17_M_realloc_insertIJRKS1_EEEvN9__gnu_cxx17__normal_iteratorIPS1_S3_EEDpOT_
.Ltmp37:
.LBB2_42:                               # %_ZNSt6vectorIPPbSaIS1_EE9push_backERKS1_.exit
                                        #   in Loop: Header=BB2_5 Depth=1
	movq	-128(%rbp), %rdi
.Ltmp38:
	callq	hipHostFree
.Ltmp39:
# %bb.43:                               #   in Loop: Header=BB2_5 Depth=1
	movq	-80(%rbp), %rdi
.Ltmp40:
	xorl	%esi, %esi
	callq	hipEventRecord
.Ltmp41:
# %bb.44:                               #   in Loop: Header=BB2_5 Depth=1
	movq	-80(%rbp), %rdi
.Ltmp42:
	callq	hipEventSynchronize
.Ltmp43:
# %bb.45:                               #   in Loop: Header=BB2_5 Depth=1
	movq	-136(%rbp), %rsi
	movq	-80(%rbp), %rdx
.Ltmp45:
	leaq	-68(%rbp), %rdi
	callq	hipEventElapsedTime
.Ltmp46:
# %bb.46:                               #   in Loop: Header=BB2_5 Depth=1
	movq	-152(%rbp), %rsi
	cmpq	-144(%rbp), %rsi
	je	.LBB2_48
# %bb.47:                               #   in Loop: Header=BB2_5 Depth=1
	movss	-68(%rbp), %xmm0                # xmm0 = mem[0],zero,zero,zero
	movss	%xmm0, -56(%rbp)                # 4-byte Spill
	movss	%xmm0, (%rsi)
	addq	$4, %rsi
	movq	%rsi, -152(%rbp)
	jmp	.LBB2_50
.LBB2_48:                               #   in Loop: Header=BB2_5 Depth=1
.Ltmp47:
	leaq	-160(%rbp), %rdi
	leaq	-68(%rbp), %rdx
	callq	_ZNSt6vectorIfSaIfEE17_M_realloc_insertIJRKfEEEvN9__gnu_cxx17__normal_iteratorIPfS1_EEDpOT_
.Ltmp48:
# %bb.49:                               # %._ZNSt6vectorIfSaIfEE9push_backERKf.exit_crit_edge
                                        #   in Loop: Header=BB2_5 Depth=1
	movss	-68(%rbp), %xmm0                # xmm0 = mem[0],zero,zero,zero
	movss	%xmm0, -56(%rbp)                # 4-byte Spill
.LBB2_50:                               # %_ZNSt6vectorIfSaIfEE9push_backERKf.exit
                                        #   in Loop: Header=BB2_5 Depth=1
	movq	-128(%rbp), %rdi
.Ltmp49:
	callq	hipHostFree
.Ltmp50:
# %bb.51:                               #   in Loop: Header=BB2_5 Depth=1
	movq	%r13, %rsp
	movss	-48(%rbp), %xmm0                # 4-byte Reload
                                        # xmm0 = mem[0],zero,zero,zero
	addss	-56(%rbp), %xmm0                # 4-byte Folded Reload
	movss	%xmm0, -48(%rbp)                # 4-byte Spill
	movq	layer_compress(%rip), %rcx
	movl	num_layers(%rip), %eax
.LBB2_52:                               #   in Loop: Header=BB2_5 Depth=1
	incq	%rbx
	movslq	%eax, %rdx
	cmpq	%rdx, %rbx
	jl	.LBB2_5
# %bb.53:                               # %.preheader260
	movq	-160(%rbp), %rax
	cmpq	%rax, -152(%rbp)
	je	.LBB2_60
# %bb.54:                               # %.lr.ph290.preheader
	xorl	%r14d, %r14d
.LBB2_55:                               # %.lr.ph290
                                        # =>This Inner Loop Header: Depth=1
	xorps	%xmm0, %xmm0
	cvtss2sd	(%rax,%r14,4), %xmm0
.Ltmp52:
	movl	$_ZSt4cout, %edi
	callq	_ZNSo9_M_insertIdEERSoT_
.Ltmp53:
# %bb.56:                               # %_ZNSolsEf.exit174
                                        #   in Loop: Header=BB2_55 Depth=1
	movq	%rax, %rbx
	movq	(%rax), %rax
	movq	-24(%rax), %rdi
	addq	%rbx, %rdi
.Ltmp54:
	movl	$10, %esi
	callq	_ZNKSt9basic_iosIcSt11char_traitsIcEE5widenEc
.Ltmp55:
# %bb.57:                               # %.noexc224
                                        #   in Loop: Header=BB2_55 Depth=1
.Ltmp56:
	movsbl	%al, %esi
	movq	%rbx, %rdi
	callq	_ZNSo3putEc
.Ltmp57:
# %bb.58:                               # %.noexc225
                                        #   in Loop: Header=BB2_55 Depth=1
.Ltmp58:
	movq	%rax, %rdi
	callq	_ZNSo5flushEv
.Ltmp59:
# %bb.59:                               # %_ZNSolsEPFRSoS_E.exit176
                                        #   in Loop: Header=BB2_55 Depth=1
	incq	%r14
	movq	-160(%rbp), %rax
	movq	-152(%rbp), %rcx
	subq	%rax, %rcx
	sarq	$2, %rcx
	cmpq	%r14, %rcx
	ja	.LBB2_55
.LBB2_60:
	movss	-48(%rbp), %xmm0                # 4-byte Reload
                                        # xmm0 = mem[0],zero,zero,zero
	jmp	.LBB2_63
.LBB2_61:                               # %._crit_edge.thread
	movaps	%xmm0, -160(%rbp)
	movq	$0, -144(%rbp)
.LBB2_62:
	xorps	%xmm0, %xmm0
.LBB2_63:                               # %._crit_edge291
	cvtss2sd	%xmm0, %xmm0
.Ltmp61:
	movl	$_ZSt4cout, %edi
	callq	_ZNSo9_M_insertIdEERSoT_
.Ltmp62:
# %bb.64:                               # %_ZNSolsEf.exit
	movq	%rax, %rbx
	movq	(%rax), %rax
	movq	-24(%rax), %rdi
	addq	%rbx, %rdi
.Ltmp63:
	movl	$10, %esi
	callq	_ZNKSt9basic_iosIcSt11char_traitsIcEE5widenEc
.Ltmp64:
# %bb.65:                               # %.noexc228
.Ltmp65:
	movsbl	%al, %esi
	movq	%rbx, %rdi
	callq	_ZNSo3putEc
.Ltmp66:
# %bb.66:                               # %.noexc229
.Ltmp67:
	movq	%rax, %rdi
	callq	_ZNSo5flushEv
.Ltmp68:
# %bb.67:                               # %_ZNSolsEPFRSoS_E.exit178.preheader
	movl	num_layers(%rip), %esi
	xorl	%edi, %edi
	testl	%esi, %esi
	jle	.LBB2_80
# %bb.68:                               # %.lr.ph306
	movq	layer_sizes(%rip), %rax
	movq	%rax, -48(%rbp)                 # 8-byte Spill
	movq	layer_compress(%rip), %r9
	movq	-208(%rbp), %rax
	movq	%rax, -168(%rbp)                # 8-byte Spill
	movl	n_threads(%rip), %r11d
	movslq	%r11d, %rax
	movq	%rax, -56(%rbp)                 # 8-byte Spill
	movl	n_compressed_data_batches(%rip), %r15d
	movslq	%r15d, %rax
	movq	%rax, -88(%rbp)                 # 8-byte Spill
	xorl	%r13d, %r13d
	xorl	%ebx, %ebx
.LBB2_69:                               # =>This Loop Header: Depth=1
                                        #     Child Loop BB2_72 Depth 2
                                        #       Child Loop BB2_74 Depth 3
	movq	-48(%rbp), %rax                 # 8-byte Reload
	movq	(%rax,%r13,8), %rcx
	cmpb	$0, (%r9,%r13)
	je	.LBB2_79
# %bb.70:                               #   in Loop: Header=BB2_69 Depth=1
	movq	%rcx, %rax
	cqto
	idivq	-56(%rbp)                       # 8-byte Folded Reload
	cqto
	idivq	-88(%rbp)                       # 8-byte Folded Reload
	testl	%r11d, %r11d
	jle	.LBB2_77
# %bb.71:                               # %.preheader259.preheader
                                        #   in Loop: Header=BB2_69 Depth=1
	movslq	%edi, %rdx
	movq	-168(%rbp), %r8                 # 8-byte Reload
	movq	(%r8,%rdx,8), %r10
	xorl	%r14d, %r14d
	xorl	%edx, %edx
.LBB2_72:                               # %.preheader259
                                        #   Parent Loop BB2_69 Depth=1
                                        # =>  This Loop Header: Depth=2
                                        #       Child Loop BB2_74 Depth 3
	testl	%r15d, %r15d
	jle	.LBB2_76
# %bb.73:                               # %.lr.ph294
                                        #   in Loop: Header=BB2_72 Depth=2
	movq	(%r10,%r14,8), %r12
	xorl	%r8d, %r8d
.LBB2_74:                               #   Parent Loop BB2_69 Depth=1
                                        #     Parent Loop BB2_72 Depth=2
                                        # =>    This Inner Loop Header: Depth=3
	cmpb	$0, (%r12,%r8)
	je	.LBB2_76
# %bb.75:                               #   in Loop: Header=BB2_74 Depth=3
	addq	%rax, %rdx
	incq	%r8
	cmpq	%r8, %r15
	jne	.LBB2_74
.LBB2_76:                               # %._crit_edge295
                                        #   in Loop: Header=BB2_72 Depth=2
	incq	%r14
	cmpq	%r11, %r14
	jne	.LBB2_72
	jmp	.LBB2_78
.LBB2_77:                               #   in Loop: Header=BB2_69 Depth=1
	xorl	%edx, %edx
.LBB2_78:                               # %._crit_edge301
                                        #   in Loop: Header=BB2_69 Depth=1
	leaq	31(%rcx), %rax
	testq	%rcx, %rcx
	cmovnsq	%rcx, %rax
	sarq	$5, %rax
	addq	%rdx, %rax
	incl	%edi
	movq	%rax, %rcx
.LBB2_79:                               # %_ZNSolsEPFRSoS_E.exit178
                                        #   in Loop: Header=BB2_69 Depth=1
	leaq	(%rbx,%rcx,4), %rbx
	incq	%r13
	cmpq	%rsi, %r13
	jne	.LBB2_69
	jmp	.LBB2_81
.LBB2_80:
	xorl	%ebx, %ebx
.LBB2_81:                               # %_ZNSolsEPFRSoS_E.exit178._crit_edge
.Ltmp70:
	movl	$_ZSt4cout, %edi
	movl	$.L.str.3, %esi
	movl	$28, %edx
	callq	_ZSt16__ostream_insertIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_PKS3_l
.Ltmp71:
# %bb.82:                               # %_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc.exit180
	xorps	%xmm0, %xmm0
	cvtsi2sd	%rbx, %xmm0
	mulsd	.LCPI2_1(%rip), %xmm0
.Ltmp72:
	movl	$_ZSt4cout, %edi
	callq	_ZNSo9_M_insertIdEERSoT_
.Ltmp73:
# %bb.83:                               # %_ZNSolsEd.exit
	movq	%rax, %rbx
	movq	(%rax), %rax
	movq	-24(%rax), %rdi
	addq	%rbx, %rdi
.Ltmp74:
	movl	$10, %esi
	callq	_ZNKSt9basic_iosIcSt11char_traitsIcEE5widenEc
.Ltmp75:
# %bb.84:                               # %.noexc232
.Ltmp76:
	movsbl	%al, %esi
	movq	%rbx, %rdi
	callq	_ZNSo3putEc
.Ltmp77:
# %bb.85:                               # %.noexc233
.Ltmp78:
	movq	%rax, %rdi
	callq	_ZNSo5flushEv
.Ltmp79:
# %bb.86:                               # %_ZNSolsEPFRSoS_E.exit183
.Ltmp80:
	movl	$_ZSt4cout, %ebx
	movl	$_ZSt4cout, %edi
	movl	$.L.str.4, %esi
	movl	$10, %edx
	callq	_ZSt16__ostream_insertIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_PKS3_l
.Ltmp81:
# %bb.87:                               # %_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc.exit185
	movq	_ZSt4cout(%rip), %rax
	addq	-24(%rax), %rbx
.Ltmp82:
	movq	%rbx, %rdi
	movl	$10, %esi
	callq	_ZNKSt9basic_iosIcSt11char_traitsIcEE5widenEc
.Ltmp83:
# %bb.88:                               # %.noexc236
.Ltmp84:
	movsbl	%al, %esi
	movl	$_ZSt4cout, %edi
	callq	_ZNSo3putEc
.Ltmp85:
# %bb.89:                               # %.noexc237
.Ltmp86:
	movq	%rax, %rdi
	callq	_ZNSo5flushEv
.Ltmp87:
# %bb.90:                               # %_ZNSolsEPFRSoS_E.exit187
	xorpd	%xmm0, %xmm0
	movapd	%xmm0, -128(%rbp)
	movq	$0, -112(%rbp)
	movl	num_layers(%rip), %r13d
	xorpd	%xmm0, %xmm0
	testl	%r13d, %r13d
	jle	.LBB2_127
# %bb.91:                               # %.lr.ph319
	movq	layer_compress(%rip), %rax
.LBB2_92:                               # =>This Loop Header: Depth=1
                                        #     Child Loop BB2_102 Depth 2
                                        #     Child Loop BB2_105 Depth 2
                                        #     Child Loop BB2_108 Depth 2
	leaq	-1(%r13), %rcx
	cmpb	$0, -1(%rax,%r13)
	je	.LBB2_118
# %bb.93:                               #   in Loop: Header=BB2_92 Depth=1
	movss	%xmm0, -88(%rbp)                # 4-byte Spill
	movq	layer_sizes(%rip), %rax
	movq	%rcx, -48(%rbp)                 # 8-byte Spill
	movq	(%rax,%rcx,8), %r12
	movq	-264(%rbp), %rax
	movq	-8(%rax), %rbx
	addq	$-8, %rax
	movq	-232(%rbp), %rcx
	movq	-8(%rcx), %rdx
	movq	%rdx, -56(%rbp)                 # 8-byte Spill
	addq	$-8, %rcx
	movq	-200(%rbp), %rdx
	movq	-8(%rdx), %r15
	addq	$-8, %rdx
	movq	%rax, -264(%rbp)
	movq	%rcx, -232(%rbp)
	movq	%rdx, -200(%rbp)
.Ltmp89:
	movl	$_ZSt4cout, %edi
	movl	$.L.str.2, %esi
	movl	$9, %edx
	callq	_ZSt16__ostream_insertIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_PKS3_l
.Ltmp90:
# %bb.94:                               # %_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc.exit189
                                        #   in Loop: Header=BB2_92 Depth=1
.Ltmp91:
	movl	$_ZSt4cout, %edi
	movq	-48(%rbp), %rsi                 # 8-byte Reload
                                        # kill: def $esi killed $esi killed $rsi
	callq	_ZNSolsEi
.Ltmp92:
# %bb.95:                               #   in Loop: Header=BB2_92 Depth=1
	movq	%rax, %r14
	movq	(%rax), %rax
	movq	-24(%rax), %rdi
	addq	%r14, %rdi
.Ltmp93:
	movl	$10, %esi
	callq	_ZNKSt9basic_iosIcSt11char_traitsIcEE5widenEc
.Ltmp94:
# %bb.96:                               # %.noexc240
                                        #   in Loop: Header=BB2_92 Depth=1
.Ltmp95:
	movsbl	%al, %esi
	movq	%r14, %rdi
	callq	_ZNSo3putEc
.Ltmp96:
# %bb.97:                               # %.noexc241
                                        #   in Loop: Header=BB2_92 Depth=1
.Ltmp97:
	movq	%rax, %rdi
	callq	_ZNSo5flushEv
.Ltmp98:
# %bb.98:                               # %_ZNSolsEPFRSoS_E.exit191
                                        #   in Loop: Header=BB2_92 Depth=1
	movq	-136(%rbp), %rdi
.Ltmp99:
	xorl	%esi, %esi
	callq	hipEventRecord
.Ltmp100:
# %bb.99:                               #   in Loop: Header=BB2_92 Depth=1
	leaq	(,%r12,4), %rsi
.Ltmp101:
	leaq	-104(%rbp), %rdi
	xorl	%edx, %edx
	callq	hipHostMalloc
.Ltmp102:
# %bb.100:                              #   in Loop: Header=BB2_92 Depth=1
	movq	%rsp, -168(%rbp)                # 8-byte Spill
	movl	n_threads(%rip), %eax
	movq	%rsp, %r14
	shlq	$4, %rax
	leaq	(%rax,%rax,2), %rax
	subq	%rax, %r14
	movq	%r14, %rsp
	movl	n_threads(%rip), %eax
	testl	%eax, %eax
	jle	.LBB2_110
# %bb.101:                              # %.lr.ph310
                                        #   in Loop: Header=BB2_92 Depth=1
	movq	-104(%rbp), %rcx
	leaq	40(%r14), %rdx
	xorl	%esi, %esi
	movq	-56(%rbp), %rdi                 # 8-byte Reload
.LBB2_102:                              #   Parent Loop BB2_92 Depth=1
                                        # =>  This Inner Loop Header: Depth=2
	movq	%rcx, -40(%rdx)
	movq	%rbx, -16(%rdx)
	movq	%r15, -8(%rdx)
	movq	%rdi, (%rdx)
	movl	%esi, -24(%rdx)
	movq	%r12, -32(%rdx)
	incq	%rsi
	addq	$48, %rdx
	cmpq	%rsi, %rax
	jne	.LBB2_102
# %bb.103:                              # %.preheader258
                                        #   in Loop: Header=BB2_92 Depth=1
	testl	%eax, %eax
	jle	.LBB2_110
# %bb.104:                              # %.lr.ph312.preheader
                                        #   in Loop: Header=BB2_92 Depth=1
	movq	%r13, -288(%rbp)                # 8-byte Spill
	movq	-64(%rbp), %r12                 # 8-byte Reload
	xorl	%r13d, %r13d
.LBB2_105:                              # %.lr.ph312
                                        #   Parent Loop BB2_92 Depth=1
                                        # =>  This Inner Loop Header: Depth=2
	movl	$_Z16decompressThreadPv, %edx
	movq	%r12, %rdi
	xorl	%esi, %esi
	movq	%r14, %rcx
	callq	pthread_create
	incq	%r13
	movslq	n_threads(%rip), %rax
	addq	$48, %r14
	addq	$8, %r12
	cmpq	%rax, %r13
	jl	.LBB2_105
# %bb.106:                              # %.preheader257
                                        #   in Loop: Header=BB2_92 Depth=1
	testl	%eax, %eax
	movq	-288(%rbp), %r13                # 8-byte Reload
	jle	.LBB2_110
# %bb.107:                              # %.lr.ph314.preheader
                                        #   in Loop: Header=BB2_92 Depth=1
	xorl	%r14d, %r14d
.LBB2_108:                              # %.lr.ph314
                                        #   Parent Loop BB2_92 Depth=1
                                        # =>  This Inner Loop Header: Depth=2
	movq	-64(%rbp), %rax                 # 8-byte Reload
	movq	(%rax,%r14,8), %rdi
.Ltmp103:
	xorl	%esi, %esi
	callq	pthread_join
.Ltmp104:
# %bb.109:                              #   in Loop: Header=BB2_108 Depth=2
	incq	%r14
	movslq	n_threads(%rip), %rax
	cmpq	%rax, %r14
	jl	.LBB2_108
.LBB2_110:                              # %._crit_edge315
                                        #   in Loop: Header=BB2_92 Depth=1
.Ltmp106:
	movq	%r15, %rdi
	callq	hipHostFree
.Ltmp107:
# %bb.111:                              #   in Loop: Header=BB2_92 Depth=1
.Ltmp108:
	movq	%rbx, %rdi
	callq	hipHostFree
.Ltmp109:
# %bb.112:                              #   in Loop: Header=BB2_92 Depth=1
.Ltmp110:
	movq	-56(%rbp), %rdi                 # 8-byte Reload
	callq	hipHostFree
.Ltmp111:
# %bb.113:                              #   in Loop: Header=BB2_92 Depth=1
	movq	-80(%rbp), %rdi
.Ltmp112:
	xorl	%esi, %esi
	callq	hipEventRecord
.Ltmp113:
# %bb.114:                              #   in Loop: Header=BB2_92 Depth=1
	movq	-80(%rbp), %rdi
.Ltmp114:
	callq	hipEventSynchronize
.Ltmp115:
# %bb.115:                              #   in Loop: Header=BB2_92 Depth=1
	movq	-136(%rbp), %rsi
	movq	-80(%rbp), %rdx
.Ltmp117:
	leaq	-96(%rbp), %rbx
	movq	%rbx, %rdi
	callq	hipEventElapsedTime
.Ltmp118:
# %bb.116:                              #   in Loop: Header=BB2_92 Depth=1
	movq	-128(%rbp), %rsi
.Ltmp120:
	leaq	-128(%rbp), %rdi
	movq	%rbx, %rdx
	callq	_ZNSt6vectorIfSaIfEE6insertEN9__gnu_cxx17__normal_iteratorIPKfS1_EERS4_
.Ltmp121:
# %bb.117:                              #   in Loop: Header=BB2_92 Depth=1
	movq	-168(%rbp), %rsp                # 8-byte Reload
	movss	-88(%rbp), %xmm0                # 4-byte Reload
                                        # xmm0 = mem[0],zero,zero,zero
	addss	-96(%rbp), %xmm0
	movq	layer_compress(%rip), %rax
	movq	-48(%rbp), %rcx                 # 8-byte Reload
.LBB2_118:                              #   in Loop: Header=BB2_92 Depth=1
	cmpq	$1, %r13
	movq	%rcx, %r13
	jg	.LBB2_92
# %bb.119:                              # %.preheader
	movq	-128(%rbp), %rax
	cmpq	%rax, -120(%rbp)
	je	.LBB2_127
# %bb.120:                              # %.lr.ph322.preheader
	movss	%xmm0, -88(%rbp)                # 4-byte Spill
	xorl	%r14d, %r14d
.LBB2_121:                              # %.lr.ph322
                                        # =>This Inner Loop Header: Depth=1
	xorps	%xmm0, %xmm0
	cvtss2sd	(%rax,%r14,4), %xmm0
.Ltmp123:
	movl	$_ZSt4cout, %edi
	callq	_ZNSo9_M_insertIdEERSoT_
.Ltmp124:
# %bb.122:                              # %_ZNSolsEf.exit195
                                        #   in Loop: Header=BB2_121 Depth=1
	movq	%rax, %rbx
	movq	(%rax), %rax
	movq	-24(%rax), %rdi
	addq	%rbx, %rdi
.Ltmp125:
	movl	$10, %esi
	callq	_ZNKSt9basic_iosIcSt11char_traitsIcEE5widenEc
.Ltmp126:
# %bb.123:                              # %.noexc244
                                        #   in Loop: Header=BB2_121 Depth=1
.Ltmp127:
	movsbl	%al, %esi
	movq	%rbx, %rdi
	callq	_ZNSo3putEc
.Ltmp128:
# %bb.124:                              # %.noexc245
                                        #   in Loop: Header=BB2_121 Depth=1
.Ltmp129:
	movq	%rax, %rdi
	callq	_ZNSo5flushEv
.Ltmp130:
# %bb.125:                              # %_ZNSolsEPFRSoS_E.exit197
                                        #   in Loop: Header=BB2_121 Depth=1
	incq	%r14
	movq	-128(%rbp), %rax
	movq	-120(%rbp), %rcx
	subq	%rax, %rcx
	sarq	$2, %rcx
	cmpq	%r14, %rcx
	ja	.LBB2_121
# %bb.126:
	movss	-88(%rbp), %xmm0                # 4-byte Reload
                                        # xmm0 = mem[0],zero,zero,zero
.LBB2_127:                              # %._crit_edge323
	cvtss2sd	%xmm0, %xmm0
.Ltmp132:
	movl	$_ZSt4cout, %edi
	callq	_ZNSo9_M_insertIdEERSoT_
.Ltmp133:
# %bb.128:                              # %_ZNSolsEf.exit193
	movq	%rax, %rbx
	movq	(%rax), %rax
	movq	-24(%rax), %rdi
	addq	%rbx, %rdi
.Ltmp134:
	movl	$10, %esi
	callq	_ZNKSt9basic_iosIcSt11char_traitsIcEE5widenEc
.Ltmp135:
# %bb.129:                              # %.noexc248
.Ltmp136:
	movsbl	%al, %esi
	movq	%rbx, %rdi
	callq	_ZNSo3putEc
.Ltmp137:
# %bb.130:                              # %.noexc249
.Ltmp138:
	movq	%rax, %rdi
	callq	_ZNSo5flushEv
.Ltmp139:
# %bb.131:                              # %_ZNSolsEPFRSoS_E.exit199.preheader
	movl	num_layers(%rip), %eax
	testl	%eax, %eax
	jle	.LBB2_135
# %bb.132:                              # %.lr.ph326
	movq	layer_sizes(%rip), %rcx
	xorl	%edx, %edx
	xorl	%esi, %esi
.LBB2_133:                              # %_ZNSolsEPFRSoS_E.exit199
                                        # =>This Inner Loop Header: Depth=1
	movq	(%rcx,%rdx,8), %rdi
	leaq	(%rsi,%rdi,4), %rsi
	incq	%rdx
	cmpq	%rdx, %rax
	jne	.LBB2_133
# %bb.134:                              # %_ZNSolsEPFRSoS_E.exit199._crit_edge.loopexit
	xorps	%xmm0, %xmm0
	cvtsi2sd	%rsi, %xmm0
	mulsd	.LCPI2_1(%rip), %xmm0
	jmp	.LBB2_136
.LBB2_135:
	xorps	%xmm0, %xmm0
.LBB2_136:                              # %_ZNSolsEPFRSoS_E.exit199._crit_edge
	movsd	%xmm0, -64(%rbp)                # 8-byte Spill
.Ltmp140:
	movl	$_ZSt4cout, %edi
	movl	$.L.str.5, %esi
	movl	$17, %edx
	callq	_ZSt16__ostream_insertIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_PKS3_l
.Ltmp141:
# %bb.137:                              # %_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc.exit201
.Ltmp142:
	movl	$_ZSt4cout, %edi
	movsd	-64(%rbp), %xmm0                # 8-byte Reload
                                        # xmm0 = mem[0],zero
	callq	_ZNSo9_M_insertIdEERSoT_
.Ltmp143:
# %bb.138:                              # %_ZNSolsEd.exit203
	movq	%rax, %rbx
	movq	(%rax), %rax
	movq	-24(%rax), %rdi
	addq	%rbx, %rdi
.Ltmp144:
	movl	$10, %esi
	callq	_ZNKSt9basic_iosIcSt11char_traitsIcEE5widenEc
.Ltmp145:
# %bb.139:                              # %.noexc252
.Ltmp146:
	movsbl	%al, %esi
	movq	%rbx, %rdi
	callq	_ZNSo3putEc
.Ltmp147:
# %bb.140:                              # %.noexc253
.Ltmp148:
	movq	%rax, %rdi
	callq	_ZNSo5flushEv
.Ltmp149:
# %bb.141:                              # %_ZNSolsEPFRSoS_E.exit205
	movq	-128(%rbp), %rdi
	testq	%rdi, %rdi
	je	.LBB2_143
# %bb.142:
	callq	_ZdlPv
.LBB2_143:                              # %_ZNSt6vectorIfSaIfEED2Ev.exit
	movq	-160(%rbp), %rdi
	testq	%rdi, %rdi
	je	.LBB2_145
# %bb.144:
	callq	_ZdlPv
.LBB2_145:                              # %_ZNSt6vectorIfSaIfEED2Ev.exit207
	movq	-280(%rbp), %rsp                # 8-byte Reload
	movq	-208(%rbp), %rdi
	testq	%rdi, %rdi
	je	.LBB2_147
# %bb.146:
	callq	_ZdlPv
.LBB2_147:                              # %_ZNSt6vectorIPPbSaIS1_EED2Ev.exit
	movq	-240(%rbp), %rdi
	testq	%rdi, %rdi
	je	.LBB2_149
# %bb.148:
	callq	_ZdlPv
.LBB2_149:                              # %_ZNSt6vectorIPjSaIS0_EED2Ev.exit
	movq	-272(%rbp), %rdi
	testq	%rdi, %rdi
	je	.LBB2_151
# %bb.150:
	callq	_ZdlPv
.LBB2_151:                              # %_ZNSt6vectorIPPPfSaIS2_EED2Ev.exit
	xorl	%eax, %eax
	leaq	-40(%rbp), %rsp
	popq	%rbx
	popq	%r12
	popq	%r13
	popq	%r14
	popq	%r15
	popq	%rbp
	.cfi_def_cfa %rsp, 8
	retq
.LBB2_152:
	.cfi_def_cfa %rbp, 16
.Ltmp6:
	movl	$_ZSt4cout, %edi
	movl	$.L.str.1, %esi
	movl	$29, %edx
	callq	_ZSt16__ostream_insertIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_PKS3_l
.Ltmp7:
# %bb.153:                              # %_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc.exit162
.Ltmp8:
	movl	$_ZSt4cout, %edi
	callq	_ZSt4endlIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_
.Ltmp9:
	jmp	.LBB2_156
.LBB2_154:
.Ltmp2:
	movl	$_ZSt4cout, %edi
	movl	$.L.str, %esi
	movl	$21, %edx
	callq	_ZSt16__ostream_insertIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_PKS3_l
.Ltmp3:
# %bb.155:                              # %_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc.exit
.Ltmp4:
	movl	$_ZSt4cout, %edi
	callq	_ZSt4endlIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_
.Ltmp5:
.LBB2_156:                              # %_ZNSolsEPFRSoS_E.exit
	xorl	%edi, %edi
	callq	exit
.LBB2_157:
.Ltmp69:
	jmp	.LBB2_169
.LBB2_158:
.Ltmp150:
	jmp	.LBB2_171
.LBB2_159:
.Ltmp88:
	jmp	.LBB2_169
.LBB2_160:                              # %.loopexit.split-lp
.Ltmp10:
	jmp	.LBB2_169
.LBB2_161:
.Ltmp119:
	jmp	.LBB2_171
.LBB2_162:
.Ltmp122:
	jmp	.LBB2_171
.LBB2_163:
.Ltmp51:
	jmp	.LBB2_169
.LBB2_164:
.Ltmp60:
	jmp	.LBB2_169
.LBB2_165:
.Ltmp131:
	jmp	.LBB2_171
.LBB2_166:
.Ltmp31:
	jmp	.LBB2_169
.LBB2_167:
.Ltmp105:
	jmp	.LBB2_171
.LBB2_168:                              # %.loopexit
.Ltmp44:
.LBB2_169:
	movq	%rax, %rbx
	jmp	.LBB2_173
.LBB2_170:
.Ltmp116:
.LBB2_171:
	movq	%rax, %rbx
	movq	-128(%rbp), %rdi
	testq	%rdi, %rdi
	je	.LBB2_173
# %bb.172:
	callq	_ZdlPv
.LBB2_173:
	movq	-160(%rbp), %rdi
	testq	%rdi, %rdi
	je	.LBB2_175
# %bb.174:
	callq	_ZdlPv
.LBB2_175:                              # %_ZNSt6vectorIfSaIfEED2Ev.exit214
	movq	-208(%rbp), %rdi
	testq	%rdi, %rdi
	je	.LBB2_177
# %bb.176:
	callq	_ZdlPv
.LBB2_177:                              # %_ZNSt6vectorIPPbSaIS1_EED2Ev.exit216
	movq	-240(%rbp), %rdi
	testq	%rdi, %rdi
	je	.LBB2_179
# %bb.178:
	callq	_ZdlPv
.LBB2_179:                              # %_ZNSt6vectorIPjSaIS0_EED2Ev.exit218
	movq	-272(%rbp), %rdi
	testq	%rdi, %rdi
	je	.LBB2_181
# %bb.180:
	callq	_ZdlPv
.LBB2_181:                              # %_ZNSt6vectorIPPPfSaIS2_EED2Ev.exit220
	movq	%rbx, %rdi
	callq	_Unwind_Resume@PLT
.Lfunc_end2:
	.size	main, .Lfunc_end2-main
	.cfi_endproc
	.section	.gcc_except_table,"a",@progbits
	.p2align	2, 0x0
GCC_except_table2:
.Lexception0:
	.byte	255                             # @LPStart Encoding = omit
	.byte	255                             # @TType Encoding = omit
	.byte	1                               # Call site Encoding = uleb128
	.uleb128 .Lcst_end0-.Lcst_begin0
.Lcst_begin0:
	.uleb128 .Lfunc_begin0-.Lfunc_begin0    # >> Call Site 1 <<
	.uleb128 .Ltmp0-.Lfunc_begin0           #   Call between .Lfunc_begin0 and .Ltmp0
	.byte	0                               #     has no landing pad
	.byte	0                               #   On action: cleanup
	.uleb128 .Ltmp0-.Lfunc_begin0           # >> Call Site 2 <<
	.uleb128 .Ltmp28-.Ltmp0                 #   Call between .Ltmp0 and .Ltmp28
	.uleb128 .Ltmp44-.Lfunc_begin0          #     jumps to .Ltmp44
	.byte	0                               #   On action: cleanup
	.uleb128 .Ltmp29-.Lfunc_begin0          # >> Call Site 3 <<
	.uleb128 .Ltmp30-.Ltmp29                #   Call between .Ltmp29 and .Ltmp30
	.uleb128 .Ltmp31-.Lfunc_begin0          #     jumps to .Ltmp31
	.byte	0                               #   On action: cleanup
	.uleb128 .Ltmp32-.Lfunc_begin0          # >> Call Site 4 <<
	.uleb128 .Ltmp43-.Ltmp32                #   Call between .Ltmp32 and .Ltmp43
	.uleb128 .Ltmp44-.Lfunc_begin0          #     jumps to .Ltmp44
	.byte	0                               #   On action: cleanup
	.uleb128 .Ltmp45-.Lfunc_begin0          # >> Call Site 5 <<
	.uleb128 .Ltmp50-.Ltmp45                #   Call between .Ltmp45 and .Ltmp50
	.uleb128 .Ltmp51-.Lfunc_begin0          #     jumps to .Ltmp51
	.byte	0                               #   On action: cleanup
	.uleb128 .Ltmp52-.Lfunc_begin0          # >> Call Site 6 <<
	.uleb128 .Ltmp59-.Ltmp52                #   Call between .Ltmp52 and .Ltmp59
	.uleb128 .Ltmp60-.Lfunc_begin0          #     jumps to .Ltmp60
	.byte	0                               #   On action: cleanup
	.uleb128 .Ltmp61-.Lfunc_begin0          # >> Call Site 7 <<
	.uleb128 .Ltmp68-.Ltmp61                #   Call between .Ltmp61 and .Ltmp68
	.uleb128 .Ltmp69-.Lfunc_begin0          #     jumps to .Ltmp69
	.byte	0                               #   On action: cleanup
	.uleb128 .Ltmp70-.Lfunc_begin0          # >> Call Site 8 <<
	.uleb128 .Ltmp87-.Ltmp70                #   Call between .Ltmp70 and .Ltmp87
	.uleb128 .Ltmp88-.Lfunc_begin0          #     jumps to .Ltmp88
	.byte	0                               #   On action: cleanup
	.uleb128 .Ltmp89-.Lfunc_begin0          # >> Call Site 9 <<
	.uleb128 .Ltmp102-.Ltmp89               #   Call between .Ltmp89 and .Ltmp102
	.uleb128 .Ltmp116-.Lfunc_begin0         #     jumps to .Ltmp116
	.byte	0                               #   On action: cleanup
	.uleb128 .Ltmp103-.Lfunc_begin0         # >> Call Site 10 <<
	.uleb128 .Ltmp104-.Ltmp103              #   Call between .Ltmp103 and .Ltmp104
	.uleb128 .Ltmp105-.Lfunc_begin0         #     jumps to .Ltmp105
	.byte	0                               #   On action: cleanup
	.uleb128 .Ltmp106-.Lfunc_begin0         # >> Call Site 11 <<
	.uleb128 .Ltmp115-.Ltmp106              #   Call between .Ltmp106 and .Ltmp115
	.uleb128 .Ltmp116-.Lfunc_begin0         #     jumps to .Ltmp116
	.byte	0                               #   On action: cleanup
	.uleb128 .Ltmp117-.Lfunc_begin0         # >> Call Site 12 <<
	.uleb128 .Ltmp118-.Ltmp117              #   Call between .Ltmp117 and .Ltmp118
	.uleb128 .Ltmp119-.Lfunc_begin0         #     jumps to .Ltmp119
	.byte	0                               #   On action: cleanup
	.uleb128 .Ltmp120-.Lfunc_begin0         # >> Call Site 13 <<
	.uleb128 .Ltmp121-.Ltmp120              #   Call between .Ltmp120 and .Ltmp121
	.uleb128 .Ltmp122-.Lfunc_begin0         #     jumps to .Ltmp122
	.byte	0                               #   On action: cleanup
	.uleb128 .Ltmp123-.Lfunc_begin0         # >> Call Site 14 <<
	.uleb128 .Ltmp130-.Ltmp123              #   Call between .Ltmp123 and .Ltmp130
	.uleb128 .Ltmp131-.Lfunc_begin0         #     jumps to .Ltmp131
	.byte	0                               #   On action: cleanup
	.uleb128 .Ltmp132-.Lfunc_begin0         # >> Call Site 15 <<
	.uleb128 .Ltmp149-.Ltmp132              #   Call between .Ltmp132 and .Ltmp149
	.uleb128 .Ltmp150-.Lfunc_begin0         #     jumps to .Ltmp150
	.byte	0                               #   On action: cleanup
	.uleb128 .Ltmp6-.Lfunc_begin0           # >> Call Site 16 <<
	.uleb128 .Ltmp5-.Ltmp6                  #   Call between .Ltmp6 and .Ltmp5
	.uleb128 .Ltmp10-.Lfunc_begin0          #     jumps to .Ltmp10
	.byte	0                               #   On action: cleanup
	.uleb128 .Ltmp5-.Lfunc_begin0           # >> Call Site 17 <<
	.uleb128 .Lfunc_end2-.Ltmp5             #   Call between .Ltmp5 and .Lfunc_end2
	.byte	0                               #     has no landing pad
	.byte	0                               #   On action: cleanup
.Lcst_end0:
	.p2align	2, 0x0
                                        # -- End function
	.section	.text._ZNSt6vectorIfSaIfEE6insertEN9__gnu_cxx17__normal_iteratorIPKfS1_EERS4_,"axG",@progbits,_ZNSt6vectorIfSaIfEE6insertEN9__gnu_cxx17__normal_iteratorIPKfS1_EERS4_,comdat
	.weak	_ZNSt6vectorIfSaIfEE6insertEN9__gnu_cxx17__normal_iteratorIPKfS1_EERS4_ # -- Begin function _ZNSt6vectorIfSaIfEE6insertEN9__gnu_cxx17__normal_iteratorIPKfS1_EERS4_
	.p2align	1, 0x90
	.type	_ZNSt6vectorIfSaIfEE6insertEN9__gnu_cxx17__normal_iteratorIPKfS1_EERS4_,@function
_ZNSt6vectorIfSaIfEE6insertEN9__gnu_cxx17__normal_iteratorIPKfS1_EERS4_: # @_ZNSt6vectorIfSaIfEE6insertEN9__gnu_cxx17__normal_iteratorIPKfS1_EERS4_
	.cfi_startproc
# %bb.0:
	pushq	%r15
	.cfi_def_cfa_offset 16
	pushq	%r14
	.cfi_def_cfa_offset 24
	pushq	%rbx
	.cfi_def_cfa_offset 32
	subq	$16, %rsp
	.cfi_def_cfa_offset 48
	.cfi_offset %rbx, -32
	.cfi_offset %r14, -24
	.cfi_offset %r15, -16
	movq	%rdi, %rbx
	movq	(%rdi), %r15
	movq	8(%rdi), %rdi
	movq	%rsi, %r14
	subq	%r15, %r14
	cmpq	16(%rbx), %rdi
	je	.LBB3_8
# %bb.1:
	cmpq	%rsi, %rdi
	je	.LBB3_2
# %bb.3:
	addq	%r14, %r15
	movss	(%rdx), %xmm1                   # xmm1 = mem[0],zero,zero,zero
	leaq	-4(%rdi), %rax
	movss	-4(%rdi), %xmm0                 # xmm0 = mem[0],zero,zero,zero
	movss	%xmm0, (%rdi)
	leaq	4(%rdi), %rcx
	movq	%rcx, 8(%rbx)
	movq	%rax, %rdx
	subq	%rsi, %rdx
	movq	%rdx, %rcx
	sarq	$2, %rcx
	cmpq	$2, %rcx
	jl	.LBB3_5
# %bb.4:
	shlq	$2, %rcx
	subq	%rcx, %rdi
	movq	%r15, %rsi
	movss	%xmm1, 12(%rsp)                 # 4-byte Spill
	callq	memmove@PLT
	movss	12(%rsp), %xmm1                 # 4-byte Reload
                                        # xmm1 = mem[0],zero,zero,zero
.LBB3_7:
	movss	%xmm1, (%r15)
	jmp	.LBB3_9
.LBB3_8:
	addq	%r14, %r15
	movq	%rbx, %rdi
	movq	%r15, %rsi
	callq	_ZNSt6vectorIfSaIfEE17_M_realloc_insertIJRKfEEEvN9__gnu_cxx17__normal_iteratorIPfS1_EEDpOT_
	jmp	.LBB3_9
.LBB3_2:
	movss	(%rdx), %xmm0                   # xmm0 = mem[0],zero,zero,zero
	movss	%xmm0, (%rsi)
	addq	$4, %rsi
	movq	%rsi, 8(%rbx)
.LBB3_9:
	addq	(%rbx), %r14
	movq	%r14, %rax
	addq	$16, %rsp
	.cfi_def_cfa_offset 32
	popq	%rbx
	.cfi_def_cfa_offset 24
	popq	%r14
	.cfi_def_cfa_offset 16
	popq	%r15
	.cfi_def_cfa_offset 8
	retq
.LBB3_5:
	.cfi_def_cfa_offset 48
	cmpq	$4, %rdx
	jne	.LBB3_7
# %bb.6:
	movss	(%r15), %xmm0                   # xmm0 = mem[0],zero,zero,zero
	movss	%xmm0, (%rax)
	jmp	.LBB3_7
.Lfunc_end3:
	.size	_ZNSt6vectorIfSaIfEE6insertEN9__gnu_cxx17__normal_iteratorIPKfS1_EERS4_, .Lfunc_end3-_ZNSt6vectorIfSaIfEE6insertEN9__gnu_cxx17__normal_iteratorIPKfS1_EERS4_
	.cfi_endproc
                                        # -- End function
	.section	.text._ZNSt6vectorIPPPfSaIS2_EE17_M_realloc_insertIJRKS2_EEEvN9__gnu_cxx17__normal_iteratorIPS2_S4_EEDpOT_,"axG",@progbits,_ZNSt6vectorIPPPfSaIS2_EE17_M_realloc_insertIJRKS2_EEEvN9__gnu_cxx17__normal_iteratorIPS2_S4_EEDpOT_,comdat
	.weak	_ZNSt6vectorIPPPfSaIS2_EE17_M_realloc_insertIJRKS2_EEEvN9__gnu_cxx17__normal_iteratorIPS2_S4_EEDpOT_ # -- Begin function _ZNSt6vectorIPPPfSaIS2_EE17_M_realloc_insertIJRKS2_EEEvN9__gnu_cxx17__normal_iteratorIPS2_S4_EEDpOT_
	.p2align	1, 0x90
	.type	_ZNSt6vectorIPPPfSaIS2_EE17_M_realloc_insertIJRKS2_EEEvN9__gnu_cxx17__normal_iteratorIPS2_S4_EEDpOT_,@function
_ZNSt6vectorIPPPfSaIS2_EE17_M_realloc_insertIJRKS2_EEEvN9__gnu_cxx17__normal_iteratorIPS2_S4_EEDpOT_: # @_ZNSt6vectorIPPPfSaIS2_EE17_M_realloc_insertIJRKS2_EEEvN9__gnu_cxx17__normal_iteratorIPS2_S4_EEDpOT_
	.cfi_startproc
# %bb.0:
	pushq	%rbp
	.cfi_def_cfa_offset 16
	pushq	%r15
	.cfi_def_cfa_offset 24
	pushq	%r14
	.cfi_def_cfa_offset 32
	pushq	%r13
	.cfi_def_cfa_offset 40
	pushq	%r12
	.cfi_def_cfa_offset 48
	pushq	%rbx
	.cfi_def_cfa_offset 56
	subq	$24, %rsp
	.cfi_def_cfa_offset 80
	.cfi_offset %rbx, -56
	.cfi_offset %r12, -48
	.cfi_offset %r13, -40
	.cfi_offset %r14, -32
	.cfi_offset %r15, -24
	.cfi_offset %rbp, -16
	movq	%rdx, (%rsp)                    # 8-byte Spill
	movq	%rsi, %r12
	movq	%rdi, %rbx
	movl	$1, %esi
	movl	$.L.str.6, %edx
	callq	_ZNKSt6vectorIPPPfSaIS2_EE12_M_check_lenEmPKc
	movq	(%rbx), %r13
	movq	8(%rbx), %r15
	movq	%r12, 8(%rsp)                   # 8-byte Spill
	subq	%r13, %r12
	movq	%r12, %r14
	sarq	$3, %r14
	testq	%rax, %rax
	movq	%rax, 16(%rsp)                  # 8-byte Spill
	je	.LBB4_1
# %bb.2:
	movq	%rbx, %rdi
	movq	%rax, %rsi
	xorl	%edx, %edx
	callq	_ZNSt15__new_allocatorIPPPfE8allocateEmPKv
	movq	%rax, %rbp
	jmp	.LBB4_3
.LBB4_1:
	xorl	%ebp, %ebp
.LBB4_3:                                # %_ZNSt12_Vector_baseIPPPfSaIS2_EE11_M_allocateEm.exit
	movq	(%rsp), %rax                    # 8-byte Reload
	movq	(%rax), %rax
	movq	%rax, (%rbp,%r14,8)
	testq	%r12, %r12
	jle	.LBB4_5
# %bb.4:
	movq	%rbp, %rdi
	movq	%r13, %rsi
	movq	%r12, %rdx
	callq	memmove@PLT
.LBB4_5:                                # %_ZNSt6vectorIPPPfSaIS2_EE11_S_relocateEPS2_S5_S5_RS3_.exit
	addq	%rbp, %r12
	addq	$8, %r12
	movq	8(%rsp), %rsi                   # 8-byte Reload
	subq	%rsi, %r15
	testq	%r15, %r15
	jle	.LBB4_7
# %bb.6:
	movq	%r12, %rdi
	movq	%r15, %rdx
	callq	memmove@PLT
.LBB4_7:                                # %_ZNSt6vectorIPPPfSaIS2_EE11_S_relocateEPS2_S5_S5_RS3_.exit16
	testq	%r13, %r13
	je	.LBB4_9
# %bb.8:
	movq	%r13, %rdi
	callq	_ZdlPv
.LBB4_9:                                # %_ZNSt12_Vector_baseIPPPfSaIS2_EE13_M_deallocateEPS2_m.exit
	addq	%r15, %r12
	movq	%rbp, (%rbx)
	movq	%r12, 8(%rbx)
	movq	16(%rsp), %rax                  # 8-byte Reload
	leaq	(,%rax,8), %rax
	addq	%rbp, %rax
	movq	%rax, 16(%rbx)
	addq	$24, %rsp
	.cfi_def_cfa_offset 56
	popq	%rbx
	.cfi_def_cfa_offset 48
	popq	%r12
	.cfi_def_cfa_offset 40
	popq	%r13
	.cfi_def_cfa_offset 32
	popq	%r14
	.cfi_def_cfa_offset 24
	popq	%r15
	.cfi_def_cfa_offset 16
	popq	%rbp
	.cfi_def_cfa_offset 8
	retq
.Lfunc_end4:
	.size	_ZNSt6vectorIPPPfSaIS2_EE17_M_realloc_insertIJRKS2_EEEvN9__gnu_cxx17__normal_iteratorIPS2_S4_EEDpOT_, .Lfunc_end4-_ZNSt6vectorIPPPfSaIS2_EE17_M_realloc_insertIJRKS2_EEEvN9__gnu_cxx17__normal_iteratorIPS2_S4_EEDpOT_
	.cfi_endproc
                                        # -- End function
	.section	.text._ZNKSt6vectorIPPPfSaIS2_EE12_M_check_lenEmPKc,"axG",@progbits,_ZNKSt6vectorIPPPfSaIS2_EE12_M_check_lenEmPKc,comdat
	.weak	_ZNKSt6vectorIPPPfSaIS2_EE12_M_check_lenEmPKc # -- Begin function _ZNKSt6vectorIPPPfSaIS2_EE12_M_check_lenEmPKc
	.p2align	1, 0x90
	.type	_ZNKSt6vectorIPPPfSaIS2_EE12_M_check_lenEmPKc,@function
_ZNKSt6vectorIPPPfSaIS2_EE12_M_check_lenEmPKc: # @_ZNKSt6vectorIPPPfSaIS2_EE12_M_check_lenEmPKc
	.cfi_startproc
# %bb.0:
	movabsq	$1152921504606846975, %rcx      # imm = 0xFFFFFFFFFFFFFFF
	movq	8(%rdi), %r8
	subq	(%rdi), %r8
	sarq	$3, %r8
	movq	%rcx, %rax
	subq	%r8, %rax
	cmpq	%rsi, %rax
	jb	.LBB5_2
# %bb.1:
	cmpq	%rsi, %r8
	cmovaq	%r8, %rsi
	leaq	(%rsi,%r8), %rax
	cmpq	%rcx, %rax
	cmovaeq	%rcx, %rax
	addq	%r8, %rsi
	cmovbq	%rcx, %rax
	retq
.LBB5_2:
	pushq	%rax
	.cfi_def_cfa_offset 16
	movq	%rdx, %rdi
	callq	_ZSt20__throw_length_errorPKc
.Lfunc_end5:
	.size	_ZNKSt6vectorIPPPfSaIS2_EE12_M_check_lenEmPKc, .Lfunc_end5-_ZNKSt6vectorIPPPfSaIS2_EE12_M_check_lenEmPKc
	.cfi_endproc
                                        # -- End function
	.section	.text._ZNSt15__new_allocatorIPPPfE8allocateEmPKv,"axG",@progbits,_ZNSt15__new_allocatorIPPPfE8allocateEmPKv,comdat
	.weak	_ZNSt15__new_allocatorIPPPfE8allocateEmPKv # -- Begin function _ZNSt15__new_allocatorIPPPfE8allocateEmPKv
	.p2align	1, 0x90
	.type	_ZNSt15__new_allocatorIPPPfE8allocateEmPKv,@function
_ZNSt15__new_allocatorIPPPfE8allocateEmPKv: # @_ZNSt15__new_allocatorIPPPfE8allocateEmPKv
	.cfi_startproc
# %bb.0:
	pushq	%rax
	.cfi_def_cfa_offset 16
	movq	%rsi, %rax
	shrq	$60, %rax
	jne	.LBB6_1
# %bb.3:
	shlq	$3, %rsi
	movq	%rsi, %rdi
	popq	%rax
	.cfi_def_cfa_offset 8
	jmp	_Znwm                           # TAILCALL
.LBB6_1:
	.cfi_def_cfa_offset 16
	shrq	$61, %rsi
	je	.LBB6_2
# %bb.4:
	callq	_ZSt28__throw_bad_array_new_lengthv
.LBB6_2:
	callq	_ZSt17__throw_bad_allocv
.Lfunc_end6:
	.size	_ZNSt15__new_allocatorIPPPfE8allocateEmPKv, .Lfunc_end6-_ZNSt15__new_allocatorIPPPfE8allocateEmPKv
	.cfi_endproc
                                        # -- End function
	.section	.text._ZNSt6vectorIPjSaIS0_EE17_M_realloc_insertIJRKS0_EEEvN9__gnu_cxx17__normal_iteratorIPS0_S2_EEDpOT_,"axG",@progbits,_ZNSt6vectorIPjSaIS0_EE17_M_realloc_insertIJRKS0_EEEvN9__gnu_cxx17__normal_iteratorIPS0_S2_EEDpOT_,comdat
	.weak	_ZNSt6vectorIPjSaIS0_EE17_M_realloc_insertIJRKS0_EEEvN9__gnu_cxx17__normal_iteratorIPS0_S2_EEDpOT_ # -- Begin function _ZNSt6vectorIPjSaIS0_EE17_M_realloc_insertIJRKS0_EEEvN9__gnu_cxx17__normal_iteratorIPS0_S2_EEDpOT_
	.p2align	1, 0x90
	.type	_ZNSt6vectorIPjSaIS0_EE17_M_realloc_insertIJRKS0_EEEvN9__gnu_cxx17__normal_iteratorIPS0_S2_EEDpOT_,@function
_ZNSt6vectorIPjSaIS0_EE17_M_realloc_insertIJRKS0_EEEvN9__gnu_cxx17__normal_iteratorIPS0_S2_EEDpOT_: # @_ZNSt6vectorIPjSaIS0_EE17_M_realloc_insertIJRKS0_EEEvN9__gnu_cxx17__normal_iteratorIPS0_S2_EEDpOT_
	.cfi_startproc
# %bb.0:
	pushq	%rbp
	.cfi_def_cfa_offset 16
	pushq	%r15
	.cfi_def_cfa_offset 24
	pushq	%r14
	.cfi_def_cfa_offset 32
	pushq	%r13
	.cfi_def_cfa_offset 40
	pushq	%r12
	.cfi_def_cfa_offset 48
	pushq	%rbx
	.cfi_def_cfa_offset 56
	subq	$24, %rsp
	.cfi_def_cfa_offset 80
	.cfi_offset %rbx, -56
	.cfi_offset %r12, -48
	.cfi_offset %r13, -40
	.cfi_offset %r14, -32
	.cfi_offset %r15, -24
	.cfi_offset %rbp, -16
	movq	%rdx, (%rsp)                    # 8-byte Spill
	movq	%rsi, %r12
	movq	%rdi, %rbx
	movl	$1, %esi
	movl	$.L.str.6, %edx
	callq	_ZNKSt6vectorIPjSaIS0_EE12_M_check_lenEmPKc
	movq	(%rbx), %r13
	movq	8(%rbx), %r15
	movq	%r12, 8(%rsp)                   # 8-byte Spill
	subq	%r13, %r12
	movq	%r12, %r14
	sarq	$3, %r14
	testq	%rax, %rax
	movq	%rax, 16(%rsp)                  # 8-byte Spill
	je	.LBB7_1
# %bb.2:
	movq	%rbx, %rdi
	movq	%rax, %rsi
	xorl	%edx, %edx
	callq	_ZNSt15__new_allocatorIPjE8allocateEmPKv
	movq	%rax, %rbp
	jmp	.LBB7_3
.LBB7_1:
	xorl	%ebp, %ebp
.LBB7_3:                                # %_ZNSt12_Vector_baseIPjSaIS0_EE11_M_allocateEm.exit
	movq	(%rsp), %rax                    # 8-byte Reload
	movq	(%rax), %rax
	movq	%rax, (%rbp,%r14,8)
	testq	%r12, %r12
	jle	.LBB7_5
# %bb.4:
	movq	%rbp, %rdi
	movq	%r13, %rsi
	movq	%r12, %rdx
	callq	memmove@PLT
.LBB7_5:                                # %_ZNSt6vectorIPjSaIS0_EE11_S_relocateEPS0_S3_S3_RS1_.exit
	addq	%rbp, %r12
	addq	$8, %r12
	movq	8(%rsp), %rsi                   # 8-byte Reload
	subq	%rsi, %r15
	testq	%r15, %r15
	jle	.LBB7_7
# %bb.6:
	movq	%r12, %rdi
	movq	%r15, %rdx
	callq	memmove@PLT
.LBB7_7:                                # %_ZNSt6vectorIPjSaIS0_EE11_S_relocateEPS0_S3_S3_RS1_.exit16
	testq	%r13, %r13
	je	.LBB7_9
# %bb.8:
	movq	%r13, %rdi
	callq	_ZdlPv
.LBB7_9:                                # %_ZNSt12_Vector_baseIPjSaIS0_EE13_M_deallocateEPS0_m.exit
	addq	%r15, %r12
	movq	%rbp, (%rbx)
	movq	%r12, 8(%rbx)
	movq	16(%rsp), %rax                  # 8-byte Reload
	leaq	(,%rax,8), %rax
	addq	%rbp, %rax
	movq	%rax, 16(%rbx)
	addq	$24, %rsp
	.cfi_def_cfa_offset 56
	popq	%rbx
	.cfi_def_cfa_offset 48
	popq	%r12
	.cfi_def_cfa_offset 40
	popq	%r13
	.cfi_def_cfa_offset 32
	popq	%r14
	.cfi_def_cfa_offset 24
	popq	%r15
	.cfi_def_cfa_offset 16
	popq	%rbp
	.cfi_def_cfa_offset 8
	retq
.Lfunc_end7:
	.size	_ZNSt6vectorIPjSaIS0_EE17_M_realloc_insertIJRKS0_EEEvN9__gnu_cxx17__normal_iteratorIPS0_S2_EEDpOT_, .Lfunc_end7-_ZNSt6vectorIPjSaIS0_EE17_M_realloc_insertIJRKS0_EEEvN9__gnu_cxx17__normal_iteratorIPS0_S2_EEDpOT_
	.cfi_endproc
                                        # -- End function
	.section	.text._ZNKSt6vectorIPjSaIS0_EE12_M_check_lenEmPKc,"axG",@progbits,_ZNKSt6vectorIPjSaIS0_EE12_M_check_lenEmPKc,comdat
	.weak	_ZNKSt6vectorIPjSaIS0_EE12_M_check_lenEmPKc # -- Begin function _ZNKSt6vectorIPjSaIS0_EE12_M_check_lenEmPKc
	.p2align	1, 0x90
	.type	_ZNKSt6vectorIPjSaIS0_EE12_M_check_lenEmPKc,@function
_ZNKSt6vectorIPjSaIS0_EE12_M_check_lenEmPKc: # @_ZNKSt6vectorIPjSaIS0_EE12_M_check_lenEmPKc
	.cfi_startproc
# %bb.0:
	movabsq	$1152921504606846975, %rcx      # imm = 0xFFFFFFFFFFFFFFF
	movq	8(%rdi), %r8
	subq	(%rdi), %r8
	sarq	$3, %r8
	movq	%rcx, %rax
	subq	%r8, %rax
	cmpq	%rsi, %rax
	jb	.LBB8_2
# %bb.1:
	cmpq	%rsi, %r8
	cmovaq	%r8, %rsi
	leaq	(%rsi,%r8), %rax
	cmpq	%rcx, %rax
	cmovaeq	%rcx, %rax
	addq	%r8, %rsi
	cmovbq	%rcx, %rax
	retq
.LBB8_2:
	pushq	%rax
	.cfi_def_cfa_offset 16
	movq	%rdx, %rdi
	callq	_ZSt20__throw_length_errorPKc
.Lfunc_end8:
	.size	_ZNKSt6vectorIPjSaIS0_EE12_M_check_lenEmPKc, .Lfunc_end8-_ZNKSt6vectorIPjSaIS0_EE12_M_check_lenEmPKc
	.cfi_endproc
                                        # -- End function
	.section	.text._ZNSt15__new_allocatorIPjE8allocateEmPKv,"axG",@progbits,_ZNSt15__new_allocatorIPjE8allocateEmPKv,comdat
	.weak	_ZNSt15__new_allocatorIPjE8allocateEmPKv # -- Begin function _ZNSt15__new_allocatorIPjE8allocateEmPKv
	.p2align	1, 0x90
	.type	_ZNSt15__new_allocatorIPjE8allocateEmPKv,@function
_ZNSt15__new_allocatorIPjE8allocateEmPKv: # @_ZNSt15__new_allocatorIPjE8allocateEmPKv
	.cfi_startproc
# %bb.0:
	pushq	%rax
	.cfi_def_cfa_offset 16
	movq	%rsi, %rax
	shrq	$60, %rax
	jne	.LBB9_1
# %bb.3:
	shlq	$3, %rsi
	movq	%rsi, %rdi
	popq	%rax
	.cfi_def_cfa_offset 8
	jmp	_Znwm                           # TAILCALL
.LBB9_1:
	.cfi_def_cfa_offset 16
	shrq	$61, %rsi
	je	.LBB9_2
# %bb.4:
	callq	_ZSt28__throw_bad_array_new_lengthv
.LBB9_2:
	callq	_ZSt17__throw_bad_allocv
.Lfunc_end9:
	.size	_ZNSt15__new_allocatorIPjE8allocateEmPKv, .Lfunc_end9-_ZNSt15__new_allocatorIPjE8allocateEmPKv
	.cfi_endproc
                                        # -- End function
	.section	.text._ZNSt6vectorIPPbSaIS1_EE17_M_realloc_insertIJRKS1_EEEvN9__gnu_cxx17__normal_iteratorIPS1_S3_EEDpOT_,"axG",@progbits,_ZNSt6vectorIPPbSaIS1_EE17_M_realloc_insertIJRKS1_EEEvN9__gnu_cxx17__normal_iteratorIPS1_S3_EEDpOT_,comdat
	.weak	_ZNSt6vectorIPPbSaIS1_EE17_M_realloc_insertIJRKS1_EEEvN9__gnu_cxx17__normal_iteratorIPS1_S3_EEDpOT_ # -- Begin function _ZNSt6vectorIPPbSaIS1_EE17_M_realloc_insertIJRKS1_EEEvN9__gnu_cxx17__normal_iteratorIPS1_S3_EEDpOT_
	.p2align	1, 0x90
	.type	_ZNSt6vectorIPPbSaIS1_EE17_M_realloc_insertIJRKS1_EEEvN9__gnu_cxx17__normal_iteratorIPS1_S3_EEDpOT_,@function
_ZNSt6vectorIPPbSaIS1_EE17_M_realloc_insertIJRKS1_EEEvN9__gnu_cxx17__normal_iteratorIPS1_S3_EEDpOT_: # @_ZNSt6vectorIPPbSaIS1_EE17_M_realloc_insertIJRKS1_EEEvN9__gnu_cxx17__normal_iteratorIPS1_S3_EEDpOT_
	.cfi_startproc
# %bb.0:
	pushq	%rbp
	.cfi_def_cfa_offset 16
	pushq	%r15
	.cfi_def_cfa_offset 24
	pushq	%r14
	.cfi_def_cfa_offset 32
	pushq	%r13
	.cfi_def_cfa_offset 40
	pushq	%r12
	.cfi_def_cfa_offset 48
	pushq	%rbx
	.cfi_def_cfa_offset 56
	subq	$24, %rsp
	.cfi_def_cfa_offset 80
	.cfi_offset %rbx, -56
	.cfi_offset %r12, -48
	.cfi_offset %r13, -40
	.cfi_offset %r14, -32
	.cfi_offset %r15, -24
	.cfi_offset %rbp, -16
	movq	%rdx, (%rsp)                    # 8-byte Spill
	movq	%rsi, %r12
	movq	%rdi, %rbx
	movl	$1, %esi
	movl	$.L.str.6, %edx
	callq	_ZNKSt6vectorIPPbSaIS1_EE12_M_check_lenEmPKc
	movq	(%rbx), %r13
	movq	8(%rbx), %r15
	movq	%r12, 8(%rsp)                   # 8-byte Spill
	subq	%r13, %r12
	movq	%r12, %r14
	sarq	$3, %r14
	testq	%rax, %rax
	movq	%rax, 16(%rsp)                  # 8-byte Spill
	je	.LBB10_1
# %bb.2:
	movq	%rbx, %rdi
	movq	%rax, %rsi
	xorl	%edx, %edx
	callq	_ZNSt15__new_allocatorIPPbE8allocateEmPKv
	movq	%rax, %rbp
	jmp	.LBB10_3
.LBB10_1:
	xorl	%ebp, %ebp
.LBB10_3:                               # %_ZNSt12_Vector_baseIPPbSaIS1_EE11_M_allocateEm.exit
	movq	(%rsp), %rax                    # 8-byte Reload
	movq	(%rax), %rax
	movq	%rax, (%rbp,%r14,8)
	testq	%r12, %r12
	jle	.LBB10_5
# %bb.4:
	movq	%rbp, %rdi
	movq	%r13, %rsi
	movq	%r12, %rdx
	callq	memmove@PLT
.LBB10_5:                               # %_ZNSt6vectorIPPbSaIS1_EE11_S_relocateEPS1_S4_S4_RS2_.exit
	addq	%rbp, %r12
	addq	$8, %r12
	movq	8(%rsp), %rsi                   # 8-byte Reload
	subq	%rsi, %r15
	testq	%r15, %r15
	jle	.LBB10_7
# %bb.6:
	movq	%r12, %rdi
	movq	%r15, %rdx
	callq	memmove@PLT
.LBB10_7:                               # %_ZNSt6vectorIPPbSaIS1_EE11_S_relocateEPS1_S4_S4_RS2_.exit16
	testq	%r13, %r13
	je	.LBB10_9
# %bb.8:
	movq	%r13, %rdi
	callq	_ZdlPv
.LBB10_9:                               # %_ZNSt12_Vector_baseIPPbSaIS1_EE13_M_deallocateEPS1_m.exit
	addq	%r15, %r12
	movq	%rbp, (%rbx)
	movq	%r12, 8(%rbx)
	movq	16(%rsp), %rax                  # 8-byte Reload
	leaq	(,%rax,8), %rax
	addq	%rbp, %rax
	movq	%rax, 16(%rbx)
	addq	$24, %rsp
	.cfi_def_cfa_offset 56
	popq	%rbx
	.cfi_def_cfa_offset 48
	popq	%r12
	.cfi_def_cfa_offset 40
	popq	%r13
	.cfi_def_cfa_offset 32
	popq	%r14
	.cfi_def_cfa_offset 24
	popq	%r15
	.cfi_def_cfa_offset 16
	popq	%rbp
	.cfi_def_cfa_offset 8
	retq
.Lfunc_end10:
	.size	_ZNSt6vectorIPPbSaIS1_EE17_M_realloc_insertIJRKS1_EEEvN9__gnu_cxx17__normal_iteratorIPS1_S3_EEDpOT_, .Lfunc_end10-_ZNSt6vectorIPPbSaIS1_EE17_M_realloc_insertIJRKS1_EEEvN9__gnu_cxx17__normal_iteratorIPS1_S3_EEDpOT_
	.cfi_endproc
                                        # -- End function
	.section	.text._ZNKSt6vectorIPPbSaIS1_EE12_M_check_lenEmPKc,"axG",@progbits,_ZNKSt6vectorIPPbSaIS1_EE12_M_check_lenEmPKc,comdat
	.weak	_ZNKSt6vectorIPPbSaIS1_EE12_M_check_lenEmPKc # -- Begin function _ZNKSt6vectorIPPbSaIS1_EE12_M_check_lenEmPKc
	.p2align	1, 0x90
	.type	_ZNKSt6vectorIPPbSaIS1_EE12_M_check_lenEmPKc,@function
_ZNKSt6vectorIPPbSaIS1_EE12_M_check_lenEmPKc: # @_ZNKSt6vectorIPPbSaIS1_EE12_M_check_lenEmPKc
	.cfi_startproc
# %bb.0:
	movabsq	$1152921504606846975, %rcx      # imm = 0xFFFFFFFFFFFFFFF
	movq	8(%rdi), %r8
	subq	(%rdi), %r8
	sarq	$3, %r8
	movq	%rcx, %rax
	subq	%r8, %rax
	cmpq	%rsi, %rax
	jb	.LBB11_2
# %bb.1:
	cmpq	%rsi, %r8
	cmovaq	%r8, %rsi
	leaq	(%rsi,%r8), %rax
	cmpq	%rcx, %rax
	cmovaeq	%rcx, %rax
	addq	%r8, %rsi
	cmovbq	%rcx, %rax
	retq
.LBB11_2:
	pushq	%rax
	.cfi_def_cfa_offset 16
	movq	%rdx, %rdi
	callq	_ZSt20__throw_length_errorPKc
.Lfunc_end11:
	.size	_ZNKSt6vectorIPPbSaIS1_EE12_M_check_lenEmPKc, .Lfunc_end11-_ZNKSt6vectorIPPbSaIS1_EE12_M_check_lenEmPKc
	.cfi_endproc
                                        # -- End function
	.section	.text._ZNSt15__new_allocatorIPPbE8allocateEmPKv,"axG",@progbits,_ZNSt15__new_allocatorIPPbE8allocateEmPKv,comdat
	.weak	_ZNSt15__new_allocatorIPPbE8allocateEmPKv # -- Begin function _ZNSt15__new_allocatorIPPbE8allocateEmPKv
	.p2align	1, 0x90
	.type	_ZNSt15__new_allocatorIPPbE8allocateEmPKv,@function
_ZNSt15__new_allocatorIPPbE8allocateEmPKv: # @_ZNSt15__new_allocatorIPPbE8allocateEmPKv
	.cfi_startproc
# %bb.0:
	pushq	%rax
	.cfi_def_cfa_offset 16
	movq	%rsi, %rax
	shrq	$60, %rax
	jne	.LBB12_1
# %bb.3:
	shlq	$3, %rsi
	movq	%rsi, %rdi
	popq	%rax
	.cfi_def_cfa_offset 8
	jmp	_Znwm                           # TAILCALL
.LBB12_1:
	.cfi_def_cfa_offset 16
	shrq	$61, %rsi
	je	.LBB12_2
# %bb.4:
	callq	_ZSt28__throw_bad_array_new_lengthv
.LBB12_2:
	callq	_ZSt17__throw_bad_allocv
.Lfunc_end12:
	.size	_ZNSt15__new_allocatorIPPbE8allocateEmPKv, .Lfunc_end12-_ZNSt15__new_allocatorIPPbE8allocateEmPKv
	.cfi_endproc
                                        # -- End function
	.section	.text._ZNSt6vectorIfSaIfEE17_M_realloc_insertIJRKfEEEvN9__gnu_cxx17__normal_iteratorIPfS1_EEDpOT_,"axG",@progbits,_ZNSt6vectorIfSaIfEE17_M_realloc_insertIJRKfEEEvN9__gnu_cxx17__normal_iteratorIPfS1_EEDpOT_,comdat
	.weak	_ZNSt6vectorIfSaIfEE17_M_realloc_insertIJRKfEEEvN9__gnu_cxx17__normal_iteratorIPfS1_EEDpOT_ # -- Begin function _ZNSt6vectorIfSaIfEE17_M_realloc_insertIJRKfEEEvN9__gnu_cxx17__normal_iteratorIPfS1_EEDpOT_
	.p2align	1, 0x90
	.type	_ZNSt6vectorIfSaIfEE17_M_realloc_insertIJRKfEEEvN9__gnu_cxx17__normal_iteratorIPfS1_EEDpOT_,@function
_ZNSt6vectorIfSaIfEE17_M_realloc_insertIJRKfEEEvN9__gnu_cxx17__normal_iteratorIPfS1_EEDpOT_: # @_ZNSt6vectorIfSaIfEE17_M_realloc_insertIJRKfEEEvN9__gnu_cxx17__normal_iteratorIPfS1_EEDpOT_
	.cfi_startproc
# %bb.0:
	pushq	%rbp
	.cfi_def_cfa_offset 16
	pushq	%r15
	.cfi_def_cfa_offset 24
	pushq	%r14
	.cfi_def_cfa_offset 32
	pushq	%r13
	.cfi_def_cfa_offset 40
	pushq	%r12
	.cfi_def_cfa_offset 48
	pushq	%rbx
	.cfi_def_cfa_offset 56
	subq	$24, %rsp
	.cfi_def_cfa_offset 80
	.cfi_offset %rbx, -56
	.cfi_offset %r12, -48
	.cfi_offset %r13, -40
	.cfi_offset %r14, -32
	.cfi_offset %r15, -24
	.cfi_offset %rbp, -16
	movq	%rdx, (%rsp)                    # 8-byte Spill
	movq	%rsi, %r12
	movq	%rdi, %rbx
	movl	$1, %esi
	movl	$.L.str.6, %edx
	callq	_ZNKSt6vectorIfSaIfEE12_M_check_lenEmPKc
	movq	(%rbx), %r13
	movq	8(%rbx), %r15
	movq	%r12, 8(%rsp)                   # 8-byte Spill
	subq	%r13, %r12
	movq	%r12, %r14
	sarq	$2, %r14
	testq	%rax, %rax
	movq	%rax, 16(%rsp)                  # 8-byte Spill
	je	.LBB13_1
# %bb.2:
	movq	%rbx, %rdi
	movq	%rax, %rsi
	xorl	%edx, %edx
	callq	_ZNSt15__new_allocatorIfE8allocateEmPKv
	movq	%rax, %rbp
	jmp	.LBB13_3
.LBB13_1:
	xorl	%ebp, %ebp
.LBB13_3:                               # %_ZNSt12_Vector_baseIfSaIfEE11_M_allocateEm.exit
	movq	(%rsp), %rax                    # 8-byte Reload
	movss	(%rax), %xmm0                   # xmm0 = mem[0],zero,zero,zero
	movss	%xmm0, (%rbp,%r14,4)
	testq	%r12, %r12
	jle	.LBB13_5
# %bb.4:
	movq	%rbp, %rdi
	movq	%r13, %rsi
	movq	%r12, %rdx
	callq	memmove@PLT
.LBB13_5:                               # %_ZNSt6vectorIfSaIfEE11_S_relocateEPfS2_S2_RS0_.exit
	addq	%rbp, %r12
	addq	$4, %r12
	movq	8(%rsp), %rsi                   # 8-byte Reload
	subq	%rsi, %r15
	testq	%r15, %r15
	jle	.LBB13_7
# %bb.6:
	movq	%r12, %rdi
	movq	%r15, %rdx
	callq	memmove@PLT
.LBB13_7:                               # %_ZNSt6vectorIfSaIfEE11_S_relocateEPfS2_S2_RS0_.exit16
	testq	%r13, %r13
	je	.LBB13_9
# %bb.8:
	movq	%r13, %rdi
	callq	_ZdlPv
.LBB13_9:                               # %_ZNSt12_Vector_baseIfSaIfEE13_M_deallocateEPfm.exit
	addq	%r15, %r12
	movq	%rbp, (%rbx)
	movq	%r12, 8(%rbx)
	movq	16(%rsp), %rax                  # 8-byte Reload
	leaq	(,%rax,4), %rax
	addq	%rbp, %rax
	movq	%rax, 16(%rbx)
	addq	$24, %rsp
	.cfi_def_cfa_offset 56
	popq	%rbx
	.cfi_def_cfa_offset 48
	popq	%r12
	.cfi_def_cfa_offset 40
	popq	%r13
	.cfi_def_cfa_offset 32
	popq	%r14
	.cfi_def_cfa_offset 24
	popq	%r15
	.cfi_def_cfa_offset 16
	popq	%rbp
	.cfi_def_cfa_offset 8
	retq
.Lfunc_end13:
	.size	_ZNSt6vectorIfSaIfEE17_M_realloc_insertIJRKfEEEvN9__gnu_cxx17__normal_iteratorIPfS1_EEDpOT_, .Lfunc_end13-_ZNSt6vectorIfSaIfEE17_M_realloc_insertIJRKfEEEvN9__gnu_cxx17__normal_iteratorIPfS1_EEDpOT_
	.cfi_endproc
                                        # -- End function
	.section	.text._ZNKSt6vectorIfSaIfEE12_M_check_lenEmPKc,"axG",@progbits,_ZNKSt6vectorIfSaIfEE12_M_check_lenEmPKc,comdat
	.weak	_ZNKSt6vectorIfSaIfEE12_M_check_lenEmPKc # -- Begin function _ZNKSt6vectorIfSaIfEE12_M_check_lenEmPKc
	.p2align	1, 0x90
	.type	_ZNKSt6vectorIfSaIfEE12_M_check_lenEmPKc,@function
_ZNKSt6vectorIfSaIfEE12_M_check_lenEmPKc: # @_ZNKSt6vectorIfSaIfEE12_M_check_lenEmPKc
	.cfi_startproc
# %bb.0:
	movabsq	$2305843009213693951, %rcx      # imm = 0x1FFFFFFFFFFFFFFF
	movq	8(%rdi), %r8
	subq	(%rdi), %r8
	sarq	$2, %r8
	movq	%rcx, %rax
	subq	%r8, %rax
	cmpq	%rsi, %rax
	jb	.LBB14_2
# %bb.1:
	cmpq	%rsi, %r8
	cmovaq	%r8, %rsi
	leaq	(%rsi,%r8), %rax
	cmpq	%rcx, %rax
	cmovaeq	%rcx, %rax
	addq	%r8, %rsi
	cmovbq	%rcx, %rax
	retq
.LBB14_2:
	pushq	%rax
	.cfi_def_cfa_offset 16
	movq	%rdx, %rdi
	callq	_ZSt20__throw_length_errorPKc
.Lfunc_end14:
	.size	_ZNKSt6vectorIfSaIfEE12_M_check_lenEmPKc, .Lfunc_end14-_ZNKSt6vectorIfSaIfEE12_M_check_lenEmPKc
	.cfi_endproc
                                        # -- End function
	.section	.text._ZNSt15__new_allocatorIfE8allocateEmPKv,"axG",@progbits,_ZNSt15__new_allocatorIfE8allocateEmPKv,comdat
	.weak	_ZNSt15__new_allocatorIfE8allocateEmPKv # -- Begin function _ZNSt15__new_allocatorIfE8allocateEmPKv
	.p2align	1, 0x90
	.type	_ZNSt15__new_allocatorIfE8allocateEmPKv,@function
_ZNSt15__new_allocatorIfE8allocateEmPKv: # @_ZNSt15__new_allocatorIfE8allocateEmPKv
	.cfi_startproc
# %bb.0:
	pushq	%rax
	.cfi_def_cfa_offset 16
	movq	%rsi, %rax
	shrq	$61, %rax
	jne	.LBB15_1
# %bb.3:
	shlq	$2, %rsi
	movq	%rsi, %rdi
	popq	%rax
	.cfi_def_cfa_offset 8
	jmp	_Znwm                           # TAILCALL
.LBB15_1:
	.cfi_def_cfa_offset 16
	shrq	$62, %rsi
	je	.LBB15_2
# %bb.4:
	callq	_ZSt28__throw_bad_array_new_lengthv
.LBB15_2:
	callq	_ZSt17__throw_bad_allocv
.Lfunc_end15:
	.size	_ZNSt15__new_allocatorIfE8allocateEmPKv, .Lfunc_end15-_ZNSt15__new_allocatorIfE8allocateEmPKv
	.cfi_endproc
                                        # -- End function
	.type	n_threads,@object               # @n_threads
	.data
	.globl	n_threads
	.p2align	2, 0x0
n_threads:
	.long	8                               # 0x8
	.size	n_threads, 4

	.type	n_compressed_data_batches,@object # @n_compressed_data_batches
	.globl	n_compressed_data_batches
	.p2align	2, 0x0
n_compressed_data_batches:
	.long	8                               # 0x8
	.size	n_compressed_data_batches, 4

	.type	layer_sizes_alexnet,@object     # @layer_sizes_alexnet
	.globl	layer_sizes_alexnet
	.p2align	4, 0x0
layer_sizes_alexnet:
	.quad	301056                          # 0x49800
	.quad	75264                           # 0x12600
	.quad	186624                          # 0x2d900
	.quad	43264                           # 0xa900
	.quad	59904                           # 0xea00
	.quad	59904                           # 0xea00
	.quad	43264                           # 0xa900
	.quad	9216                            # 0x2400
	.size	layer_sizes_alexnet, 64

	.type	layer_compress_alexnet,@object  # @layer_compress_alexnet
	.globl	layer_compress_alexnet
layer_compress_alexnet:
	.zero	8,1
	.size	layer_compress_alexnet, 8

	.type	layer_density_alexnet,@object   # @layer_density_alexnet
	.globl	layer_density_alexnet
	.p2align	4, 0x0
layer_density_alexnet:
	.quad	50                              # 0x32
	.quad	80                              # 0x50
	.quad	40                              # 0x28
	.quad	60                              # 0x3c
	.quad	70                              # 0x46
	.quad	70                              # 0x46
	.quad	30                              # 0x1e
	.quad	60                              # 0x3c
	.size	layer_density_alexnet, 64

	.type	num_layers_alexnet,@object      # @num_layers_alexnet
	.globl	num_layers_alexnet
	.p2align	2, 0x0
num_layers_alexnet:
	.long	8                               # 0x8
	.size	num_layers_alexnet, 4

	.type	layer_sizes_vgg,@object         # @layer_sizes_vgg
	.globl	layer_sizes_vgg
	.p2align	4, 0x0
layer_sizes_vgg:
	.quad	3211264                         # 0x310000
	.quad	3211264                         # 0x310000
	.quad	802816                          # 0xc4000
	.quad	1605632                         # 0x188000
	.quad	1605632                         # 0x188000
	.quad	401408                          # 0x62000
	.quad	802816                          # 0xc4000
	.quad	802816                          # 0xc4000
	.quad	802816                          # 0xc4000
	.quad	200704                          # 0x31000
	.quad	401408                          # 0x62000
	.quad	401408                          # 0x62000
	.quad	401408                          # 0x62000
	.quad	100352                          # 0x18800
	.quad	100352                          # 0x18800
	.quad	100352                          # 0x18800
	.quad	100352                          # 0x18800
	.quad	25088                           # 0x6200
	.size	layer_sizes_vgg, 144

	.type	layer_density_vgg,@object       # @layer_density_vgg
	.globl	layer_density_vgg
	.p2align	4, 0x0
layer_density_vgg:
	.quad	50                              # 0x32
	.quad	20                              # 0x14
	.quad	30                              # 0x1e
	.quad	20                              # 0x14
	.quad	10                              # 0xa
	.quad	20                              # 0x14
	.quad	20                              # 0x14
	.quad	20                              # 0x14
	.quad	10                              # 0xa
	.quad	20                              # 0x14
	.quad	20                              # 0x14
	.quad	10                              # 0xa
	.quad	10                              # 0xa
	.quad	10                              # 0xa
	.quad	20                              # 0x14
	.quad	20                              # 0x14
	.quad	10                              # 0xa
	.quad	15                              # 0xf
	.size	layer_density_vgg, 144

	.type	layer_compress_vgg,@object      # @layer_compress_vgg
	.globl	layer_compress_vgg
	.p2align	4, 0x0
layer_compress_vgg:
	.zero	18,1
	.size	layer_compress_vgg, 18

	.type	num_layers_vgg,@object          # @num_layers_vgg
	.globl	num_layers_vgg
	.p2align	2, 0x0
num_layers_vgg:
	.long	18                              # 0x12
	.size	num_layers_vgg, 4

	.type	layer_sizes,@object             # @layer_sizes
	.globl	layer_sizes
	.p2align	3, 0x0
layer_sizes:
	.quad	layer_sizes_alexnet
	.size	layer_sizes, 8

	.type	layer_compress,@object          # @layer_compress
	.globl	layer_compress
	.p2align	3, 0x0
layer_compress:
	.quad	layer_compress_alexnet
	.size	layer_compress, 8

	.type	layer_density,@object           # @layer_density
	.globl	layer_density
	.p2align	3, 0x0
layer_density:
	.quad	layer_density_alexnet
	.size	layer_density, 8

	.type	num_layers,@object              # @num_layers
	.globl	num_layers
	.p2align	2, 0x0
num_layers:
	.long	8                               # 0x8
	.size	num_layers, 4

	.type	.L.str,@object                  # @.str
	.section	.rodata.str1.1,"aMS",@progbits,1
.L.str:
	.asciz	"bad number of threads"
	.size	.L.str, 22

	.type	.L.str.1,@object                # @.str.1
.L.str.1:
	.asciz	"bad num_elements or n_threads"
	.size	.L.str.1, 30

	.type	.L.str.2,@object                # @.str.2
.L.str.2:
	.asciz	"starting "
	.size	.L.str.2, 10

	.type	.L.str.3,@object                # @.str.3
.L.str.3:
	.asciz	"total_space_compressed(MB): "
	.size	.L.str.3, 29

	.type	.L.str.4,@object                # @.str.4
.L.str.4:
	.asciz	"decompress"
	.size	.L.str.4, 11

	.type	.L.str.5,@object                # @.str.5
.L.str.5:
	.asciz	"total space(MB): "
	.size	.L.str.5, 18

	.type	.L.str.6,@object                # @.str.6
.L.str.6:
	.asciz	"vector::_M_realloc_insert"
	.size	.L.str.6, 26

	.type	__hip_cuid_,@object             # @__hip_cuid_
	.bss
	.globl	__hip_cuid_
__hip_cuid_:
	.byte	0                               # 0x0
	.size	__hip_cuid_, 1

	.ident	"AMD clang version 18.0.0git (https://github.com/RadeonOpenCompute/llvm-project roc-6.3.2 25012 e5bf7e55c91490b07c49d8960fa7983d864936c4)"
	.section	".note.GNU-stack","",@progbits
	.addrsig
	.addrsig_sym _Z14compressThreadPv
	.addrsig_sym _Z16decompressThreadPv
	.addrsig_sym __gxx_personality_v0
	.addrsig_sym _Unwind_Resume
	.addrsig_sym layer_sizes_alexnet
	.addrsig_sym layer_compress_alexnet
	.addrsig_sym layer_density_alexnet
	.addrsig_sym _ZSt4cout
	.addrsig_sym __hip_cuid_

	.text
	.p2alignl 7, 3214868480
	.fill 96, 4, 3214868480
	.type	__hip_cuid_,@object             ; @__hip_cuid_
	.section	.bss,"aw",@nobits
	.globl	__hip_cuid_
__hip_cuid_:
	.byte	0                               ; 0x0
	.size	__hip_cuid_, 1

	.ident	"AMD clang version 18.0.0git (https://github.com/RadeonOpenCompute/llvm-project roc-6.3.2 25012 e5bf7e55c91490b07c49d8960fa7983d864936c4)"
	.section	".note.GNU-stack","",@progbits
	.addrsig
	.addrsig_sym __hip_cuid_
	.amdgcn_target "amdgcn-amd-amdhsa--gfx1100"
	.amdgpu_metadata
---
amdhsa.kernels:  []
amdhsa.target:   amdgcn-amd-amdhsa--gfx1100
amdhsa.version:
  - 1
  - 2
...

	.end_amdgpu_metadata
