Community
cancel
Showing results for 
Search instead for 
Did you mean: 
Altera_Forum
Honored Contributor I
897 Views

Compiler Error: Multiple channel write sites.

Hi, 

 

My code as follows: 

 

#ifndef SIMD_WORK_ITEMS 

#define SIMD_WORK_ITEMS 4 // default value 

#endif 

 

 

#pragma OPENCL_EXTENSION cl_altera_channels : enable 

 

 

// Channel declarations 

channel int DATA __attribute__((depth(8))); 

 

 

__kernel 

__attribute((reqd_work_group_size(BLOCK_SIZE,BLOCK_SIZE,1))) 

__attribute((num_simd_work_items(SIMD_WORK_ITEMS))) 

void matrixMult( __global int *restrict A, __global int *restrict B) 

// Write result to channel 

int off = get_global_id(1) * get_global_size(0) + get_global_id(0); 

int running_sum = A[off] * B[off]; 

write_channel_altera(DATA, running_sum); 

 

 

__kernel 

void FindMinAndMax(int size, __global int *restrict data) 

int min_temp = 0; 

int max_temp = 0; 

 

 

for (int i = 0; i < size; i++) { 

int val = read_channel_altera(DATA); 

min_temp = min(min_temp, val); 

max_temp = max(max_temp, val); 

 

 

data[i] = val; 

 

 

When I compile the kernels with command "aoc device/matrix_mult_bak.cl -o bin/matrix_mult_bak.aocx --report -v", it print out errors  

"Compiler Error: Multiple channel write sites.". What the error means?How to solve the problem?Thank you!
0 Kudos
2 Replies
Altera_Forum
Honored Contributor I
65 Views

I have solved the problem, just delete "__attribute((num_simd_work_items(SIMD_WORK_ITEMS)) )" .

Altera_Forum
Honored Contributor I
65 Views

For future reference, "Compiler Error: Multiple channel write sites." means that you are writing to the same channel from multiple locations in your code, which is not allowed. Each channel can only have one read and one write port. When you use SIMD, the compiler will try to extend the kernel pipeline to run multiple work-items in parallel, and if you have a channel read or write in your code, this will result in multiple work-items trying to read from or write to the same channel at the same time which is not allowed. SIMD and kernel pipeline replication (num_compute_units) are not supported for NDRange kernels that have channels.

Reply