Intel® Quartus® Prime Software
Intel® Quartus® Prime Design Software, Design Entry, Synthesis, Simulation, Verification, Timing Analysis, System Design (Platform Designer, formerly Qsys)
16597 Discussions

Compiler Error: Multiple channel write sites.

Altera_Forum
Honored Contributor II
1,277 Views

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 II
445 Views

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

0 Kudos
Altera_Forum
Honored Contributor II
445 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.

0 Kudos
Reply